c++ - 使用 CUDA 在 GPU 之间复制大数据

Question

我是 CUDA 的新手，我自己正在尝试一项非常简单的任务——将内存复制到显卡，然后再复制回来。这是我编写的代码的简化版本。

const int arraySize = 100; 

int* data1 = NULL; 
int* data2 = NULL;
// Initialized both data1 and data2
// ... 

int* dev_data1 = NULL; 
int* dev_data2 = NULL; 
// Initialized both dev_data1 and dev_data2
// ... 

// copy data1 to device
cudaMemcpy(dev_data1, data1, arraySize*sizeof(int), cudaMemcpyHostToDevice );

// copy dev_data1 to dev_data2 with gpu
gpuCopy<<<1, arraySize>>>( dev_data1, dev_data2 ); 

// copy dev_data2 to data
cudaMemcpy(data2, dev_data2, arraySize*sizeof(int), cudaMemcpyDeviceToHost );

而gpuGopy如下:

__global__ void gpucopy( int* src, int* dst )
{
    int i = threadIdx.x;
    dst[i] = src[i];
} 

我发现如果 arraySize 很小，上面的函数就可以工作。但如果 arraySize 达到特定大小，data2 将变为全零。我的猜测是运行 gpu 函数时存在某种限制。但是有办法找出来吗？如果我有一个非常大的数组，我如何将它复制到 GPU(并返回)？

Answer 1

首先你应该做proper cuda error checking

其次，您可能认为 threadIdx.x 给出了一个全局唯一的线程 ID。它没有。

因此修改您的内核行:

int i = threadIdx.x;

到:

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

最后，您的每 block 线程参数不能超过 512 或 1024，具体取决于 GPU。我们通过指定每个 block 的线程数和每个网格参数的 block 数来启动更大的网格:

#define nTPB 512
gpuCopy<<<(arraySize + nTPB - 1)/nTPB, nTPB>>>( dev_data1, dev_data2 ); 

结合这种网格大小调整方法，我们通常在内核中包含线程检查，以防止对任意网格/问题大小的越界访问:

__global__ void gpucopy( int* src, int* dst, int size )
{
    int i = threadIdx.x + blockDim.x * blockIdx.x;
    if (i < size)
      dst[i] = src[i];
}

同样，我们必须告诉内核问题的大小:

gpuCopy<<<(arraySize + nTPB - 1)/nTPB, nTPB>>>( dev_data1, dev_data2, arraySize ); 

您可能想要查看 CUDA programming guide