ceph/src/boost/libs/compute/perf/perf_thrust_reduce_by_key.cu

   1 //---------------------------------------------------------------------------//
   2 // Copyright (c) 2015 Jakub Szuppe <j.szuppe@gmail.com>
   3 //
   4 // Distributed under the Boost Software License, Version 1.0
   5 // See accompanying file LICENSE_1_0.txt or copy at
   6 // http://www.boost.org/LICENSE_1_0.txt
   7 //
   8 // See http://boostorg.github.com/compute for more information.
   9 //---------------------------------------------------------------------------//
  10
  11 #include <algorithm>
  12 #include <cstdlib>
  13 #include <iostream>
  14
  15 #include <thrust/copy.h>
  16 #include <thrust/device_vector.h>
  17 #include <thrust/generate.h>
  18 #include <thrust/host_vector.h>
  19 #include <thrust/reduce.h>
  20
  21 #include "perf.hpp"
  22
  23 int rand_int()
  24 {
  25     return static_cast<int>((rand() / double(RAND_MAX)) * 25.0);
  26 }
  27
  28 struct unique_key {
  29   int current;
  30   int avgValuesNoPerKey;
  31
  32   unique_key()
  33   {
  34       current = 0;
  35       avgValuesNoPerKey = 512;
  36   }
  37
  38   int operator()()
  39   {
  40       double p = double(1.0) / static_cast<double>(avgValuesNoPerKey);
  41       if((rand() / double(RAND_MAX)) <= p)
  42           return ++current;
  43       return current;
  44   }
  45 } UniqueKey;
  46
  47 int main(int argc, char *argv[])
  48 {
  49     perf_parse_args(argc, argv);
  50
  51     std::cout << "size: " << PERF_N << std::endl;
  52
  53     // create vector of keys and random values
  54     thrust::host_vector<int> host_keys(PERF_N);
  55     thrust::host_vector<int> host_values(PERF_N);
  56     std::generate(host_keys.begin(), host_keys.end(), UniqueKey);
  57     std::generate(host_values.begin(), host_values.end(), rand_int);
  58
  59     // transfer data to the device
  60     thrust::device_vector<int> device_keys = host_keys;
  61     thrust::device_vector<int> device_values = host_values;
  62
  63     // create device vectors for the results
  64     thrust::device_vector<int> device_keys_results(PERF_N);
  65     thrust::device_vector<int> device_values_results(PERF_N);
  66
  67     typedef typename thrust::device_vector<int>::iterator iterType;
  68     thrust::pair<iterType, iterType> result;
  69
  70     perf_timer t;
  71     for(size_t trial = 0; trial < PERF_TRIALS; trial++){
  72         t.start();
  73         result = thrust::reduce_by_key(device_keys.begin(),
  74                                        device_keys.end(),
  75                                        device_values.begin(),
  76                                        device_keys_results.begin(),
  77                                        device_values_results.begin());
  78         cudaDeviceSynchronize();
  79         t.stop();
  80     }
  81     std::cout << "time: " << t.min_time() / 1e6 << " ms" << std::endl;
  82
  83     size_t result_size = thrust::distance(device_keys_results.begin(), result.first);
  84     if(result_size != static_cast<size_t>(host_keys[PERF_N-1] + 1)){
  85         std::cout << "ERROR: "
  86                   << "wrong number of keys"
  87                   << std::endl;
  88         return -1;
  89     }
  90
  91     return 0;
  92 }