如何使用 cub::DeviceReduce::ArgMin()答案

【问题标题】：How to use cub::DeviceReduce::ArgMin()如何使用 cub::DeviceReduce::ArgMin()
【发布时间】：2020-09-08 03:50:23
【问题描述】：

我对如何使用 cub::DeviceReduce::ArgMin() 有一些困惑。这里我从CUB的文档中复制代码。

#include <cub/cub.cuh> 
// Declare, allocate, and initialize device-accessible pointers for input and output
int                      num_items;      // e.g., 7
int                      *d_in;          // e.g., [8, 6, 7, 5, 3, 0, 9], located in GPU
KeyValuePair<int, int>   *d_out;         // e.g., [{-,-}]

// Determine temporary device storage requirements
void     *d_temp_storage = NULL;
size_t   temp_storage_bytes = 0;
cub::DeviceReduce::ArgMin(d_temp_storage, temp_storage_bytes, d_in, d_out, num_items);
// Allocate temporary storage
cudaMalloc(&d_temp_storage, temp_storage_bytes);
// Run argmin-reduction
cub::DeviceReduce::ArgMin(d_temp_storage, temp_storage_bytes, d_in, d_out, num_items);
// d_out <-- [{5, 0}]

问题如下：

如果d_in是指向某个GPU内存（设备）的指针，如何初始化d_out的指针？
如果 ArgMin() 的操作在设备 (GPU) 中完成，如何将结果复制到我的 CPU？

【问题讨论】：

github.com/NVlabs/cub/blob/1.8.0/examples/device/…
只包含函数cub::DeviceReduce::Sum()的用法，它们完全不同。
不，他们不是。它们基本相同，只是结果的返回类型是一对，而不是标量

标签： c++ cuda cub

【解决方案1】：

如果d_in是指向某个GPU内存（设备）的指针，如何初始化d_out的指针？

您使用cudaMalloc，类似于初始化d_in 指针的方式。

如果 ArgMin() 的操作在设备 (GPU) 中完成，我如何将结果复制到我的 CPU？

您使用cudaMemcpy，类似于您将d_in 数据从主机复制到设备的方式，但现在您将d_out 数据从设备复制到主机。 KeyValuePair 是具有 key 和 value 成员的 C++ 对象。

这是一个完整的例子：

$ cat t37.cu
#include <cub/cub.cuh>
#include <iostream>

int main(){


  // Declare, allocate, and initialize device-accessible pointers for input and output
  int                      num_items = 32;
  int                      *d_in;
  cub::KeyValuePair<int, int>   *d_out;

  int *h_in = new int[num_items];
  cub::KeyValuePair<int, int> *h_out = new cub::KeyValuePair<int, int>;
  cudaMalloc(&d_in, num_items*sizeof(d_in[0]));
  cudaMalloc(&d_out, sizeof(cub::KeyValuePair<int, int>));
  for (int i = 0; i < num_items; i++) h_in[i] = 4;
  h_in[12] = 2;  // so we expect our return tuple to be 12,2
  cudaMemcpy(d_in, h_in, num_items*sizeof(d_in[0]), cudaMemcpyHostToDevice);

  // Determine temporary device storage requirements
  void     *d_temp_storage = NULL;
  size_t   temp_storage_bytes = 0;
  cub::DeviceReduce::ArgMin(d_temp_storage, temp_storage_bytes, d_in, d_out, num_items);
  // Allocate temporary storage
  cudaMalloc(&d_temp_storage, temp_storage_bytes);
  // Run argmin-reduction
  cub::DeviceReduce::ArgMin(d_temp_storage, temp_storage_bytes, d_in, d_out, num_items);

  cudaMemcpy(h_out, d_out, sizeof(cub::KeyValuePair<int, int>), cudaMemcpyDeviceToHost);
  std::cout << "minimum value: " << h_out[0].value << std::endl;
  std::cout << "index of min:  " << h_out[0].key << std::endl;
}
$ nvcc -o t37 t37.cu -arch=sm_35 -std=c++14 -Wno-deprecated-gpu-targets
$ ./t37
minimum value: 2
index of min:  12
$

【讨论】：