为什么线程块中的变量具有相同的内存地址？库达

Question

我想知道为什么它们有相同的内存地址，如果我没记错的话，每个线程都有一个自己的创建变量的副本：

__global__ void
Matrix_Multiplication_Shared(
   const int* const Matrix_A, 
   const int* const Matrix_B, 
         int* const Matrix_C)
{   
    const int sum_value = threadIdx.x;
    printf("%p \n", &sum_value);
}

输出：

我正在考虑一个线程块的情况，例如有 2 个或更多线程。

Answer 1

NVIDIA GPU 有多个地址空间。

指针使用的主要虚拟地址空间称为通用地址空间。在通用地址空间内是本地内存和共享内存的窗口。通用地址空间的其余部分是全局地址空间。 PTX 和 GPU 指令集支持基于 0 访问本地和共享内存地址空间的附加指令。

一些自动变量和堆栈内存在本地内存地址空间中。全局内存和本地内存之间的主要区别在于，本地内存的组织方式是通过连续的线程 ID 访问连续的 32 位字。如果每个线程从相同的本地内存偏移量读取或写入，则内存访问完全合并。

在 PTX 中，本地内存通过 ld.local 和 st.local 访问。

在 GPU SASS 中，指令有两种形式：

1. LDL、STL 是对本地内存的直接访问，以基于 0 的偏移量给出
2. LD、ST 可用于通过通用本地内存窗口访问本地内存。

当您获取变量的地址时，将返回通用地址空间地址。每个线程都看到与通用本地内存窗口基指针相同的偏移量。加载存储单元会将基于 0 的偏移量转换为每个线程唯一的全局地址。

欲了解更多信息，请参阅：

• CUDA Programming Guide 本地内存部分

• PTX ISA 通用寻址部分。本地内存的详细信息散布在整个手册中。