在CUDA內核

Question

聲明變量假設你在CUDA內核聲明一個新的變量，然後在多線程使用它，比如：

__global__ void kernel(float* delt, float* deltb) { 
int i = blockIdx.x * blockDim.x + threadIdx.x; 
float a; 
a = delt[i] + deltb[i]; 
a += 1; 
} 

和內核調用看起來像下面，多線程和塊：

int threads = 200; 
uint3 blocks = make_uint3(200,1,1); 
kernel<<<blocks,threads>>>(d_delt, d_deltb); 

1. 是 「A」 存儲在堆棧上？
2. 是一個新的「一」爲每個線程創建時被初始化？
3. 或者將每個線程獨立訪問「一」在一個未知的時間，有可能搞亂的算法？

Answer 1

以上皆非。在CUDA編譯器足夠聰明，足夠有侵略性的優化，它可以檢測到a未使用，完整的代碼可以被優化away.You可以通過編譯內核-Xptxas=-v證實了這一點作爲一個選項，並期待在資源計數，這應該基本上沒有寄存器，也沒有本地內存或堆。

在一個不那麼簡單的例子中，a可能存儲在每個線程寄存器或每個線程本地存儲器中，這是脫機DRAM。

Answer 2

在內核函數中聲明的任何變量（標量或數組）都是不帶外部說明符的，對於每個線程都是本地的，也就是說每個線程都有自己的該變量的「副本」，線程之間不會發生數據競爭！

要由編譯器來選擇局部變量是否會駐留在寄存器或本地內存（實際上是全局內存），這取決於編譯器執行變換和優化。

有關哪些變量存在本地存儲器的更多詳細信息，請參見NVIDIA Cuda用戶指南第5.3.2.2章