CUDA __global__ 函数中内核参数的内存空间

Question

在如下的 CUDA 函数中：

__global__ void Kernel(int value) {
    value += 1;
    ...
}

void Host() {
    Kernel<<<10, 10>>>(123);
}

value 的内存空间是在Kernel 设备内部（全局）、共享还是本地？

如果一个线程修改了它，那么该修改对其他线程是否可见？还是变量位于每个线程的堆栈上，就像函数内部定义的变量一样？

Answer 1

内核设备内部的值的内存空间是（全局的）、共享的还是本地的？

它在逻辑本地空间中。作为内核启动过程的一部分，内核参数从特定的__constant__ 内存库开始。然而，对于大多数实际使用，参数将首先被复制到线程本地寄存器，它是逻辑本地空间的一部分。即使对于不是 LD 但可以引用__constant__ 内存的 SASS 指令，它的使用实际上是本地的，每个线程，就像寄存器是本地的，每个线程一样。

如果一个线程对其进行了修改，该修改对其他线程是否可见？

一个线程中的修改对其他线程是不可见的。如果你修改它，修改将（首先）对其在线程本地寄存器中的值执行。

或者变量位于每个线程的堆栈中，就像函数内部定义的变量一样？

堆栈位于线程的逻辑本地空间中，因此我不确定该问题的目的是什么。一个线程的堆栈不与另一个线程共享。根据我的经验，这样一个变量出现在堆栈上的唯一方法是，如果它被用作函数调用进程的一部分（即不是线程本身，因为它最初是由内核启动进程产生的，而是源自那个线程）。

此外，在函数内部定义的变量（例如局部变量）也不一定会出现在堆栈中。这主要是编译器决定的功能。它们可能在寄存器中，它们可能出现（例如由于溢出）在实际设备内存中（但仍然在逻辑本地空间中），或者它们可能在堆栈中，在某些时候，可能作为函数调用的一部分。

这应该可以使用CUDA binary utilities 进行验证。