CUDA 中的每个内核调用是不是保证唯一线程 ID？

Posted 2023-03-23

【中文标题】CUDA 中的每个内核调用是不是保证唯一线程 ID？

【英文标题】：Is Unique Thread Id guaranteed for each Kernel Call in CUDA?CUDA 中的每个内核调用是否保证唯一线程 ID？

【发布时间】：2021-03-17 22:51:39

【问题描述】：

我最近开始使用 Cuda，我有 C++、Java 和 Python 的多线程、多进程编码经验。

使用 PyCuda，我看到了这样的示例代码，

ker = SourceModule("""
__global__ void scalar_multiply_kernel(float *outvec, float scalar, float *vec)

     int i = threadIdx.x;
     outvec[i] = scalar*vec[i];

""")

似乎线程 id 本身参与了代码的逻辑。那么问题是是否有足够的线程 id 覆盖我的整个数组（我显然需要对其的索引到达那里的所有元素），以及如果我更改数组的大小会发生什么。

索引是否总是在 0 和 N 之间？

【问题讨论】：

总之没有。在编程指南docs.nvidia.com/cuda/cuda-c-programming-guide/…

的前几页中有大量文档说明这是如何工作的

【参考方案1】：

在 CUDA 中，线程 ID 仅在每个所谓的线程块中是唯一的，这意味着，您的示例内核只做正确的事情，而只有一个块在工作。这可能是在早期的示例中完成的，以使您更容易理解这些想法，但就性能而言，这样做通常是一件非常糟糕的事情：

使用一个块，您只能使用 GPU 中的多个流式多处理器 (SM) 之一，即使该 SM 在等待期间有足够的并行工作要做时，也只能隐藏内存访问延迟。

如果您的内核不包含循环，那么单个线程块也会限制您的线程数量，从而限制问题大小，因此每个线程都可以计算多个元素。

内核执行具有很强的层次性：为简单起见，我们将自己限制为一维索引，内核在所谓的gridDim.x线程块网格上执行，每个线程块包含blockDim.x线程，每个块编号为threadIdx.x，而每个块都通过blockIdx.x编号。

要获得线程的唯一 ID（以理想的方式使用硬件从数组中加载元素），您必须采用 blockIdx.x * blockDim.x + threadIdx.x。如果每个线程要计算多个元素，则使用以下形式的循环

for (int i = blockIdx.x * blockDim.x + threadIdx.x; i < InputSize; i += gridDim.x * blockDim.x)  
/* ... */

这称为网格步长循环，因为gridDim.x * blockDim.x 是在内核上工作的所有线程的数量。不同的步幅（尤其是让线程处理连续元素：步幅 = 1）可能会起作用，但由于内存访问模式不理想，会慢得多。