2012-06-30 112 views
1

具有2.x設備的設備中的銀行衝突是什麼?據我瞭解CUDA C編程指南,在2.x設備中,如果兩個線程在同一個共享內存庫中訪問相同的32位字,它不會導致銀行衝突。相反,這個詞是廣播。當兩個線程在同一個共享內存組中寫入相同的32位字時,則只有一個線程成功。由於片上存儲器爲64KB(共享內存爲48KB,L1爲16KB,反之亦然),並且它以32個存儲體組織,因此我假設每個存儲體都由2 KB組成。所以我認爲如果兩個線程在同一個共享內存區中訪問兩個不同的32位字,就會出現銀行衝突。它是否正確?2.x設備中的銀行衝突

+1

這是正確的。我有種感覺像用-1來解決你的問題,因爲在CUDA C編程指南中對銀行衝突進行了徹底的解釋。 –

回答

3

您的描述是正確的。有許多訪問模式可能會導致銀行衝突,但以下是一個簡單而常見的示例:跨接訪問。

__shared__ int smem[512]; 

int tid = threadIdx.x; 

x = smem[tid * 2]; // 2-way bank conflicts 
y = smem[tid * 4]; // 4-way bank conflicts 
z = smem[tid * 8]; // 8-way bank conflicts 
// etc. 

銀行ID =指數%32,因此,如果查看地址的在x,y中的圖案,和z訪問,則可以看到,在32個線程的每個經線,對於x,2個線程將訪問每個銀行,對於y,4個線程將訪問每個銀行,並且對於z,8個線程將訪問每個銀行。