cuda中的cuda函数应用程序元素

Question

在将矩阵 A 和向量 x 相乘得到结果 y 后，我想将函数 h 逐元素应用于 y。

我想获得 z = h(Ax)，其中 h 按元素应用于向量 Ax。

我知道如何在 GPU 上进行矩阵/向量乘法（使用 cublas）。现在我希望 h（这是我自己的函数，用 C++ 编码）也应用于 GPU 中的结果向量，我该怎么做？

Answer 1

两种可能的方法是：

1. 编写自己的 CUDA 内核来执行操作
2. 使用thrust（例如thrust::for_each()）。

以下是这两种方法的一个有效示例：

$ cat t934.cu
#include <iostream>
#include <thrust/host_vector.h>
#include <thrust/device_vector.h>
#include <thrust/copy.h>
#include <thrust/for_each.h>

#define DSIZE 4

#define nTPB 256

template <typename T>
__host__ __device__ T myfunc(T &d){

  return d + 5;  // define your own function here
}

struct mytfunc
{
template <typename T>
__host__ __device__
 void operator()(T &d){

  d = myfunc(d);
  }
};

template <typename T>
__global__ void mykernel(T *dvec, size_t dsize){

  int idx = threadIdx.x+blockDim.x*blockIdx.x;
  if (idx < dsize) dvec[idx] = myfunc(dvec[idx]);
}

int main(){

  // first using kernel
  float *h_data, *d_data;
  h_data = new float[DSIZE];
  cudaMalloc(&d_data, DSIZE*sizeof(float));
  for (int i = 0; i < DSIZE; i++) h_data[i] = i;
  cudaMemcpy(d_data, h_data, DSIZE*sizeof(float), cudaMemcpyHostToDevice);
  mykernel<<<(DSIZE+nTPB-1)/nTPB,nTPB>>>(d_data, DSIZE);
  cudaMemcpy(h_data, d_data, DSIZE*sizeof(float), cudaMemcpyDeviceToHost);
  for (int i = 0; i < DSIZE; i++) std::cout << h_data[i] << ",";
  std::cout << std::endl;

  // then using thrust
  thrust::host_vector<float>   hvec(h_data, h_data+DSIZE);
  thrust::device_vector<float> dvec = hvec;
  thrust::for_each(dvec.begin(), dvec.end(), mytfunc());
  thrust::copy_n(dvec.begin(), DSIZE, std::ostream_iterator<float>(std::cout, ","));
  std::cout << std::endl;
}

$ nvcc -o t934 t934.cu
$ ./t934
5,6,7,8,
10,11,12,13,
$

请注意，为了提供一个完整的示例，我从主机内存中的向量定义开始。如果您已经在设备内存中有向量（可能是计算 y=Ax 的结果），那么您可以直接处理它，通过将该向量传递给 CUDA 内核，或直接在推力函数中使用它，使用 @987654324 @wrapper（这个方法在之前链接的thrust快速入门指南中有介绍。）

我在这里所做的假设是您想使用一个变量的任意函数。这应该可以处理myfunc 中定义的几乎任意函数。但是，对于您可能感兴趣的某些类别的函数，您也可以通过一个或多个 CUBLAS 调用来实现它。