通过CUDA线程复制全局内存

Question

我需要通过CUDA线程(不是从主机)将全局内存中的一个数组复制到全局内存中的另一个数组。

我的代码如下：

__global__ void copy_kernel(int *g_data1, int *g_data2, int n)
{
  int idx = blockIdx.x * blockDim.x + threadIdx.x;
  int start, end;
  start = some_func(idx);
  end = another_func(idx);
  unsigned int i;
  for (i = start; i < end; i++) {
      g_data2[i] = g_data1[idx];
  }
}

这是非常低效的，因为对于某些idx，开始和结束区域非常大，这使得线程发出太多的复制命令。有没有办法有效地实现它？

谢谢,

郑

Answer 1

尝试使用以下命令：

CUresult cuMemcpyDtoD(
    CUdeviceptr dst,
    CUdeviceptr src,
    unsigned int bytes   
)   

更新：

你说得对：http://forums.nvidia.com/index.php?showtopic=88745

没有有效的方法来正确地做到这一点，因为CUDA的设计要求您在内核中只使用少量的数据。

Answer 2

按照你写的方式，我猜每个线程都在尝试写整个“start”到“end”块。这真的很低效。

你需要做这样的事情。

___shared___ unsigned sm_start[BLOCK_SIZE];
___shared___ unsigned sm_end[BLOCK_SIZE];
sm_start[threadIdx.x] = start;
sm_end[threadIdx.y] = end;
__syncthreads();
for (int n = 0; n < blockdDim.x; n++) {
g_data2 += sm_start[n];
unsigned lim = sm_end[n] - sm_start[n];
  for (int i = threadIdx.x; i < lim; i += blockDim.x) {
      g_data2[i] = g_data1[idx];
  }
}