使用内核参数在CUDA内核中声明数组

Question

我在C ++ / CUDA编程并遇到了一个问题：

__global__ void KERNEL(int max_size, double* x, double* y, double* z)
{
      double localArray_x[max_size]
      double localArray_y[max_size]
      double localArray_z[max_size]
      //do stuff here
}

现在我唯一的解决方案是预定义max_size，如下所示：

#define max_size 20

这些数组是我内核工作的主要焦点。基本上，我有全局坐标，只有这些坐标的部分（基于模拟框内的位置）被添加到三个local_arrays中。然后在这些坐标上完成工作，最后在模拟结束时将这些坐标添加回全局数组（x，y，z）。因此，对数组有一些限制：

1. 调用的每个线程都应该有max_size * 3个数组元素来操作。
2. 这些数组被广泛使用，因此内核需要能够快速（或本地）访问它们。
3. max_size不能是常量，因为我的坐标的数密度是根据主机的输入而变化的。

我知道StackOverflow上有这篇文章的版本，但我相信我需要的不同于简单的共享内存声明。我只是在寻找可以做什么以及这些选项中最快的选项的指导。

如果相关，max_size将在每次模拟中相同（恒定）。换句话说，它只会从一个模拟变为另一个模拟，而不会在同一模拟中变化。

Answer 1

这比我想象的要简单。使用new和delete来实现此目的，就像在主机上执行此操作一样。

唯一的区别是需要使用运行时API调用来为您的目的在堆上分配内存：

cudaDeviceSetLimit(cudaLimitMallocHeapSize, heapsize);

其中，对于运行N个内核的系统，其中包含3个大小为N_SIZE的int数组：

size_t heapsize = (size_t)( N*3*N_SIZE*sizeof(int) );