File: perf_thrust_reduce_by_key.cu

package info (click to toggle)
boost1.88 1.88.0-1
  • links: PTS, VCS
  • area: main
  • in suites: forky, sid, trixie
  • size: 576,932 kB
  • sloc: cpp: 4,149,234; xml: 136,789; ansic: 35,092; python: 33,910; asm: 5,698; sh: 4,604; ada: 1,681; makefile: 1,633; pascal: 1,139; perl: 1,124; sql: 640; yacc: 478; ruby: 271; java: 77; lisp: 24; csh: 6
file content (92 lines) | stat: -rw-r--r-- 2,788 bytes parent folder | download | duplicates (14)
1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
25
26
27
28
29
30
31
32
33
34
35
36
37
38
39
40
41
42
43
44
45
46
47
48
49
50
51
52
53
54
55
56
57
58
59
60
61
62
63
64
65
66
67
68
69
70
71
72
73
74
75
76
77
78
79
80
81
82
83
84
85
86
87
88
89
90
91
92
//---------------------------------------------------------------------------//
// Copyright (c) 2015 Jakub Szuppe <j.szuppe@gmail.com>
//
// Distributed under the Boost Software License, Version 1.0
// See accompanying file LICENSE_1_0.txt or copy at
// http://www.boost.org/LICENSE_1_0.txt
//
// See http://boostorg.github.com/compute for more information.
//---------------------------------------------------------------------------//

#include <algorithm>
#include <cstdlib>
#include <iostream>

#include <thrust/copy.h>
#include <thrust/device_vector.h>
#include <thrust/generate.h>
#include <thrust/host_vector.h>
#include <thrust/reduce.h>

#include "perf.hpp"

int rand_int()
{
    return static_cast<int>((rand() / double(RAND_MAX)) * 25.0);
}

struct unique_key {
  int current;
  int avgValuesNoPerKey;

  unique_key()
  {
      current = 0;
      avgValuesNoPerKey = 512;
  }

  int operator()()
  {
      double p = double(1.0) / static_cast<double>(avgValuesNoPerKey);
      if((rand() / double(RAND_MAX)) <= p)
          return ++current;
      return current;
  }
} UniqueKey;

int main(int argc, char *argv[])
{
    perf_parse_args(argc, argv);

    std::cout << "size: " << PERF_N << std::endl;
    
    // create vector of keys and random values
    thrust::host_vector<int> host_keys(PERF_N);
    thrust::host_vector<int> host_values(PERF_N);
    std::generate(host_keys.begin(), host_keys.end(), UniqueKey);
    std::generate(host_values.begin(), host_values.end(), rand_int);
    
    // transfer data to the device
    thrust::device_vector<int> device_keys = host_keys;
    thrust::device_vector<int> device_values = host_values;

    // create device vectors for the results
    thrust::device_vector<int> device_keys_results(PERF_N);
    thrust::device_vector<int> device_values_results(PERF_N);

    typedef typename thrust::device_vector<int>::iterator iterType;
    thrust::pair<iterType, iterType> result;

    perf_timer t;
    for(size_t trial = 0; trial < PERF_TRIALS; trial++){
        t.start();
        result = thrust::reduce_by_key(device_keys.begin(),
                                       device_keys.end(),
                                       device_values.begin(),
                                       device_keys_results.begin(),
                                       device_values_results.begin());
        cudaDeviceSynchronize();
        t.stop();
    }
    std::cout << "time: " << t.min_time() / 1e6 << " ms" << std::endl;
    
    size_t result_size = thrust::distance(device_keys_results.begin(), result.first);
    if(result_size != static_cast<size_t>(host_keys[PERF_N-1] + 1)){
        std::cout << "ERROR: "
                  << "wrong number of keys"
                  << std::endl;
        return -1;
    }

    return 0;
}