在 nvidia cuda 内核中创建数组答案

【问题标题】：creating arrays in nvidia cuda kernel在 nvidia cuda 内核中创建数组
【发布时间】：2011-01-12 07:57:52
【问题描述】：

嗨，我只是想知道是否可以在 nvidia cuda 内核中执行以下操作

__global__ void compute(long *c1, long size, ...)
{
  ...
  long d[1000];
  ...
}

或以下

__global__ void compute(long *c1, long size, ...)
{
  ...
  long d[size];
  ...
}

【问题讨论】：

标签： c++ cuda nvidia

【解决方案1】：

第一个例子你可以做，第二个我没试过。

但是，如果您能提供帮助，您可能希望重新设计您的程序以不这样做。您不想在内核中分配 4000 字节的内存。这将导致大量使用 CUDA 本地内存，因为您无法将所有内容都放入寄存器中。 CUDA 本地内存很慢（400 个循环的内存延迟）。

【讨论】：

这个问题的答案组合很有价值。特别要注意 Sebastian Dressler 的观点，即每个线程都将分配这 4000 个字节。因此，如果您启动一个由 1024 个块组成的网格，每个块包含 256 个线程，则需要大约 1GB 的内存。

【解决方案2】：

您可以执行 #1，但请注意这将在每个线程中完成！

第二个 sn-p 不起作用，因为不支持内核运行时的动态内存分配。

【讨论】：

【解决方案3】：

您可以在启动内核时动态分配共享内存。

__global__ void compute(long *c1, long size, ...)
 {
  ...
   extern __shared__ float shared[];
  ...
 }

compute <<< dimGrid, dimBlock, sharedMemSize >>>( blah blah );

CUDA 编程指南：

数组的大小在启动时确定（参见第 4.2.3 节）。

【讨论】：

【解决方案4】：

支持内核运行时动态内存分配，查看 sdk 示例，新删除。

【讨论】：