CUDA:如何为类的数据成员分配内存

Question

假设我有这门课程:

class Particle
{
    double *_w;
};

我想将nParticles对象发送Particle到我的内核.为这些对象分配空间很简单:

Particle *dev_p;
cudaStatus = cudaMalloc((void**)&dev_P, nParticles * sizeof(Particle));
if (cudaStatus != cudaSuccess) {
    fprintf(stderr, "cudaMalloc failed!");
    goto Error;
}

还假设 nParticles是100.现在,我需要分配300双为每_w一个Particle对象.我怎样才能做到这一点？我试过这段代码:

for( int i = 0; i < nParticles; i++){
    cudaStatus = cudaMalloc((void**)&(dev_P[i]._w), 300 * sizeof(double));
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "cudaMalloc failed!");
        goto Error;
    }
}

但是当我访问dev_p [i] ._ w [j]时,使用Nsight进行调试会停止.

Answer 1

也许你应该包括一个完整的简单例子.(如果我在上面编译你的代码并自己运行它,在linux上,我在第二次cudaMalloc操作时遇到了一个seg错误).我看到的一个问题是,由于您在第一步中已经在设备内存中分配了粒子对象,因此当您分配_w指针时,您将指向已经在设备内存中的cudaMalloc.您应该将基于主机的指针传递给cudaMalloc,然后它将分配给设备(全局)内存中的已分配区域.

我认为一个可能的解决方案符合我在yoru示例中看到的内容是这样的:

#include <stdio.h>

#define cudaCheckErrors(msg) \
    do { \
        cudaError_t __err = cudaGetLastError(); \
        if (__err != cudaSuccess) { \
            fprintf(stderr, "Fatal error: %s (%s at %s:%d)\n", \
                msg, cudaGetErrorString(__err), \
                __FILE__, __LINE__); \
            fprintf(stderr, "*** FAILED - ABORTING\n"); \
            exit(1); \
        } \
    } while (0)

class Particle
{
    public:
    double *_w;
};

__global__ void test(Particle *p){

  int idx=threadIdx.x + blockDim.x*blockIdx.x;

  if (idx == 2){
    printf("dev_p[2]._w[2] = %f\n", p[idx]._w[2]);
    }
}


int main() {
  int nParticles=100;
  Particle *dev_p;
  double *w[nParticles];
  cudaMalloc((void**)&dev_p, nParticles * sizeof(Particle));
  cudaCheckErrors("cudaMalloc1 fail");

  for( int i = 0; i < nParticles; i++){
    cudaMalloc((void**)&(w[i]), 300 * sizeof(double));
    cudaCheckErrors("cudaMalloc2 fail");
    cudaMemcpy(&(dev_p[i]._w), &(w[i]), sizeof(double *), cudaMemcpyHostToDevice);
    cudaCheckErrors("cudaMemcpy1 fail");
    }
  double testval = 32.7;
  cudaMemcpy(w[2]+2, &testval, sizeof(double), cudaMemcpyHostToDevice);
  cudaCheckErrors("cudaMemcpy2 fail");
  test<<<1, 32>>>(dev_p);
  cudaDeviceSynchronize();
  cudaCheckErrors("kernel fail");
  printf("Done!\n");

}

这里我们在主机上创建一组单独的指针用于cudaMalloc目的,然后将那些分配的指针复制到设备以用作设备指针(这对于UVA来说是合法的).

另一种方法是在设备端分配_w指针.这也可能符合您的目的.

以上所有我假设cc 2.0或更高.