在 elementwise add() CUDA 内核中，为什么步幅为 blockDim.x * gridDim.x？

Question

我正在学习 CUDA 编程，但我无法理解这个元素加法内核中的步幅：

// performs vector addition
// a, b, c are vectors and added values are stored in a and b, while the results are stored in c.

#define N 10

__global__ void add(int* a, int* b, int* c) {
    int tid = threadIdx.x + blockIdx.x*blockDim.x;
    while(tid < N) {
        c[tid] = a[tid] + b[tid];
        tid += blockDim.x * gridDim.x;
    }
}

这是这个的 CPU 版本：

void add(int* a, int* b, int* c) {
    int tid = 0;    // start from CPU 0
    while(tid < N) {
        c[tid] = a[tid] + b[tid];
        tid += 2;   // assume there are 2 CPUs
    }
}

我浏览了几个教程，所有教程都是从这个 add 内核开始的。

不明白tid的步幅从何而来？

编辑

现在我有点明白步幅值是什么意思了。我认为这意味着如果我有 2 CPUs，当两个运行线程之一完成时，我必须将 tid 添加 2因为 tid + 1 由另一个核心处理。

但是问题来了，CPU1中的tid和CPU2中的tid有区别吗？我认为它们不能是同一个变量，并且值存储在 CPUs?

的单独内存中

Answer 1

在CUDA编程模型中，计算由"threads"的"blocks"执行。每个线程在块内都有一个线程 ID 和一个块 ID。因此，就像您使用 CPU 的示例一样，如果您使用 2 个大小为 3 的块启动内核，您将有 6 个线程：

块 0，线程 0
块 0，线程 1
块 0，线程 2
块 1，线程 0
块 1，线程 1
块 1，线程 2

tid 变量将是网格中的整体线程 ID，结合了块和块内线程 ID。在示例中，请注意这些线程的 tid 值将完全覆盖范围 0..5。

现在，如果 6 个线程各自对数组元素 0..5 执行添加操作，现在想继续处理其他元素，我们可以推进它们的索引以确保它们中的每一个都获得一个新的和不同的索引继续工作，并且没有发现任何元素。步幅为 6（在我们的示例中）这样做：线程将处理 6..11、12..17、18..23 等等。所以第一个线程将在 0、6、12、18 等上工作，第二个线程将在 1、7、13、19 等上工作，依此类推。

内核写成下面这样或许会更清晰：

__global__ void add(int* a, int* b, int* c) {
    int overall_thread_id = threadIdx.x + blockIdx.x*blockDim.x;
    int overall_num_threads = blockDim.x * gridDim.x;
    int pos = overall_thread_id;
    while(pos < N) {
        c[pos] = a[pos] + b[pos];
        pos += overall_num_threads;
    }
}

至于你的 "CPU version" 代码 - 它不会工作，因为即使你有不同的线程（可能在不同的内核上）执行它，它们都会从 tid 开始为 0并以相同的方式前进——不像 GPU "threads"，每个都以不同的 tid 开始。如果你有一个 CPU 函数初始化 tid = index_of_thread_among_Workers() 和 2 个工作线程，第一个线程将在 0、2、4、6 等上工作，第二个工作线程（索引为 1）将在 1 上工作, 3, 5, 7 等等

在 elementwise add() CUDA 内核中，为什么步幅为 blockDim.x * gridDim.x？

In an elementwise add() CUDA kernel, why is the stride blockDim.x * gridDim.x?

cuda

gpgpu