只能将一个类传递给CUDA内核进行并行评估？

我有一个相当复杂的C++类，它具有类myObj.fun(x,y,z)的功能。我想在一个3D网格点上的GPU上调用它。只能将一个类传递给CUDA内核进行并行评估？

我的高层次的问题是：通过myObj和大量的点到GPU是一件简单的事情吗？由于工作原因，我避免创建此函数的CUDA实现，但对我而言，这可能非常简单。

同样，这是一个非常高层次的问题，所以“是的，这很容易”或“不，它更复杂”是受欢迎的，尽管有点方向也会有所帮助。

2016-07-25 drjrm3

可以在GPU上使用类/对象，包括它们的方法（例如fun()）。这样的类至少必须具有用__host__ __device__修饰的方法，但代码重构可能不会比这更多地涉及。然而，这样的方法（像没有任何重构的其他代码一样）可能不会访问GPU的任何并行功能。最基本的比较是，在单个CPU线程中运行的方法会在单个GPU线程中运行。这通常不会更快，如果您只是将单个对象传递给GPU并在GPU上运行相同的单线程代码（在单个GPU线程中），则速度通常会更慢。

一种可能的策略是，如果你有很多这些对象，或者在你的情况下有相当多的“点”，代表要独立完成的工作，那么你可以通过每一个（对象或点）到GPU线程，并以这种方式处理它们，以便实现GPU所喜欢的大规模多线程操作。理想情况下，你将有10,000或更多的点来处理这种方式。

由于各种原因，这仍然不是最有效的GPU使用方式，其中一个原因与高效的数据访问有关，另一个原因与（可能的）线程分歧有关。尽管如此，有些人确实追求这种“简单”，“不同寻常的并行”的代码移植方法，偶尔也会有趣的加速。

根据您的实际代码，如果您以允许相邻线程访问相邻数据的方式将点传递给GPU，对于访问点的每个操作，您可能会看到有吸引力的结果。期望你可以以这种方式获得有吸引力的加速是相当合理的，可能只需要相对较少的代码重构，但要注意数据组织以实现最佳GPU访问。

这里是一个完全样例：

$ cat t30.cu 
#include <iostream> 
#include <cstdlib> 

const int dsize = 3; 
const int nTPB = 256; 
const int rng = 8; 

class myclass 
{ 

    int increment; 
    public: 
    myclass(int _incr): increment(_incr) {}; 
    // methods callable on the device need the __device__ decoration 
    __host__ __device__ void fun(int &x, int &y, int &z){ 
     x += increment; 
     y += increment; 
     z += increment;} 

}; 

// this is the actual device routine that is run per thread 
__global__ void mykernel(myclass obj, int *dx, int *dy, int *dz, int dsize){ 

    int idx = threadIdx.x+blockDim.x*blockIdx.x; // figure out which thread we are 
    if (idx < dsize) 
    obj.fun(dx[idx], dy[idx], dz[idx]); // apply method 
} 


int main(){ 

    // allocate host data 
    int *p_x, *p_y, *p_z, *d_x, *d_y, *d_z; 
    p_x = new int[dsize]; 
    p_y = new int[dsize]; 
    p_z = new int[dsize]; 

    // allocate device data 
    cudaMalloc(&d_x, dsize*sizeof(int)); 
    cudaMalloc(&d_y, dsize*sizeof(int)); 
    cudaMalloc(&d_z, dsize*sizeof(int)); 

    // initialize host data 
    std::cout << "Before:" << std::endl; 
    for (int i = 0; i < dsize; i++){ 
    p_x[i] = rand()%rng; 
    p_y[i] = rand()%rng; 
    p_z[i] = rand()%rng; 
    std::cout << p_x[i] << "," << p_y[i] << "," << p_z[i] << std::endl;} 

    // copy to device 
    cudaMemcpy(d_x, p_x, dsize*sizeof(int), cudaMemcpyHostToDevice); 
    cudaMemcpy(d_y, p_y, dsize*sizeof(int), cudaMemcpyHostToDevice); 
    cudaMemcpy(d_z, p_z, dsize*sizeof(int), cudaMemcpyHostToDevice); 

    // instantiate object on host 
    myclass test(1); 

    // copy object to device as kernel parameter 
    mykernel<<<(dsize+nTPB-1)/nTPB, nTPB>>>(test, d_x, d_y, d_z, dsize); 

    // copy data back to host 
    cudaMemcpy(p_x, d_x, dsize*sizeof(int), cudaMemcpyDeviceToHost); 
    cudaMemcpy(p_y, d_y, dsize*sizeof(int), cudaMemcpyDeviceToHost); 
    cudaMemcpy(p_z, d_z, dsize*sizeof(int), cudaMemcpyDeviceToHost); 


    std::cout << "After:" << std::endl; 
    for (int i = 0; i < dsize; i++){ 
    std::cout << p_x[i] << "," << p_y[i] << "," << p_z[i] << std::endl;} 

    return 0; 
} 
$ nvcc -o t30 t30.cu 
$ ./t30 
Before: 
7,6,1 
3,1,7 
2,4,1 
After: 
8,7,2 
4,2,8 
3,5,2 
$

为了表达简洁，我省略了proper cuda error checking但我总是建议你使用它时，你正在开发CUDA代码。

来源

2016-07-25 01:26:19

只能将一个类传递给CUDA内核进行并行评估？

回答

相关问题