關於CUDA的bank conflict - 顯卡

By Sandy
at 2009-06-18T17:39
at 2009-06-18T17:39
Table of Contents
※ 引述《BeelZeBub (天使獵人)》之銘言:
: ※ 引述《rick209 ()》之銘言:
: : 近來閱讀了版上a大關於cuda的文章
: : 因此想把自己的程式改寫成可利用GPU執行
: : 但bank conflict卻始終困擾著我
: : 舉例來說
: : for(int k=0; k<num; k++){
: : num1=data1[k]
: : num2=data2[k]
: : sum[k]=num1+num2;
: : }
: : 若把num1 num2的記憶體配置在shared memory時
: : 會因為不同執行緒存取到同一塊記憶體產生bank conflict的問題
: : 但因為計算複雜 所需記憶體大的關係 也無法配置到暫存器上
: : 想請教cuda有類似openMp中 for private()的指令嗎
: : 還是就只能完全利用陣列運算 如把num1改變成陣列num1[k]等
: 來討論一下
: CUDA的threading的確會有memory conflict的問提存在
: 假設今天開四個threading
: [th0 th0 th0 th1 th1 th1 th2 th2 th2 th3 th3 th3]
: 根據memory allocation
: 每個thread的第一個job(1st th0, 1st th1.....)都會同時卡在第一塊block裡面
: 這樣會造成bottleneck 而使的performance上不來
: 以上如果我沒理解錯應該就是你的問題吧
: 但是今天不要像上面規劃的那樣 規劃成
: [th0 th1 th2 th3 th0 th1 th2 th3 th0 th1 th2 th3]
: 如此一來 每個thread的第一個job就可以錯開同一block的存取
: 概念是這樣 如果要寫成程式的話
: 就要改指向頭的header (講index比較好)
: 詳細怎樣作要想一想 不過概念應該是這樣沒錯
: 歡迎版上大大討論... @@!
抱歉 寫得有點錯誤 自己在補充一下
int k=threadIdx.x
num1=data1[k]
num2=data2[k]
sum[k]=num1+num2;
因為我的程式中 需要大量的迴圈
因此最簡單的平行方式 就是每個thread負責一個運算
k=1 thread 1負責
k=2 thread 2負責
依此類推
但是同一個block中 共用共享記憶體
因此有可能
thread1在存取num1時 thread2也剛好存取num1
如此會造成最終運算結果錯誤
由於變數眾多 因此需將變數配置於shared memory中
感謝B大跟D大的回文
需要時間消化一下
--
: ※ 引述《rick209 ()》之銘言:
: : 近來閱讀了版上a大關於cuda的文章
: : 因此想把自己的程式改寫成可利用GPU執行
: : 但bank conflict卻始終困擾著我
: : 舉例來說
: : for(int k=0; k<num; k++){
: : num1=data1[k]
: : num2=data2[k]
: : sum[k]=num1+num2;
: : }
: : 若把num1 num2的記憶體配置在shared memory時
: : 會因為不同執行緒存取到同一塊記憶體產生bank conflict的問題
: : 但因為計算複雜 所需記憶體大的關係 也無法配置到暫存器上
: : 想請教cuda有類似openMp中 for private()的指令嗎
: : 還是就只能完全利用陣列運算 如把num1改變成陣列num1[k]等
: 來討論一下
: CUDA的threading的確會有memory conflict的問提存在
: 假設今天開四個threading
: [th0 th0 th0 th1 th1 th1 th2 th2 th2 th3 th3 th3]
: 根據memory allocation
: 每個thread的第一個job(1st th0, 1st th1.....)都會同時卡在第一塊block裡面
: 這樣會造成bottleneck 而使的performance上不來
: 以上如果我沒理解錯應該就是你的問題吧
: 但是今天不要像上面規劃的那樣 規劃成
: [th0 th1 th2 th3 th0 th1 th2 th3 th0 th1 th2 th3]
: 如此一來 每個thread的第一個job就可以錯開同一block的存取
: 概念是這樣 如果要寫成程式的話
: 就要改指向頭的header (講index比較好)
: 詳細怎樣作要想一想 不過概念應該是這樣沒錯
: 歡迎版上大大討論... @@!
抱歉 寫得有點錯誤 自己在補充一下
int k=threadIdx.x
num1=data1[k]
num2=data2[k]
sum[k]=num1+num2;
因為我的程式中 需要大量的迴圈
因此最簡單的平行方式 就是每個thread負責一個運算
k=1 thread 1負責
k=2 thread 2負責
依此類推
但是同一個block中 共用共享記憶體
因此有可能
thread1在存取num1時 thread2也剛好存取num1
如此會造成最終運算結果錯誤
由於變數眾多 因此需將變數配置於shared memory中
感謝B大跟D大的回文
需要時間消化一下
--
Tags:
顯卡
All Comments

By Odelette
at 2009-06-20T10:28
at 2009-06-20T10:28

By Dorothy
at 2009-06-25T01:35
at 2009-06-25T01:35

By John
at 2009-06-29T14:43
at 2009-06-29T14:43

By Ula
at 2009-07-03T16:37
at 2009-07-03T16:37

By Necoo
at 2009-07-04T11:27
at 2009-07-04T11:27
Related Posts
3850 開1920*1080 使用hdmi>dvi線

By Damian
at 2009-06-18T17:13
at 2009-06-18T17:13
老電腦想升級顯示卡

By Jack
at 2009-06-18T16:07
at 2009-06-18T16:07
關於CUDA的bank conflict

By Susan
at 2009-06-18T16:03
at 2009-06-18T16:03
9600GT VS 3850

By Yuri
at 2009-06-18T15:42
at 2009-06-18T15:42
該用哪張顯卡比較好呢?

By Megan
at 2009-06-18T14:47
at 2009-06-18T14:47