Re: [問題] CUDA新手上路 平行度及硬體相關問題

作者: friends29 (涼哥哥)   2016-12-03 17:38:53
※ 引述《tmbyksdG (雨神妹妹的男朋友)》之銘言:
: 開發平台(Platform): (Ex: Win10, Linux, ...)
: Linux上安裝CUDA環境 (CUDA版本為8.0 運算能力為3.7)(Tesla K80)
: 編譯器(Ex: GCC, clang, VC++...)+目標環境(跟開發平台不同的話需列出)
: NVCC
: 額外使用到的函數庫(Library Used): (Ex: OpenGL, ...)
: 問題(Question):
: 硬體方面:
: 1.我執行deviceQuery偵測到2個device(device0:Tesla K80, device1:K80),估狗發現K80
: 是由兩個GK210核心所組成,那我偵測到的device是指有兩個K80(4個GK210)的意思嗎?
: 還是偵測到的兩個device其實就是GK210?
: 2.13個SMX,總共有2496個cores,所以我一次可以同時做運算的數量是否為2496個threads?
: 軟體方面:
: 我寫了一個64 * 64的矩陣乘法,我想測試不同的block & thread數量去做運算,哪個執
: 行時間會比較短,我試了兩種block的配置(thread數量剛好為4096,一個thread執行一
: 個輸出矩陣的一個element)。
: (1)dim3 dimBlock(32, 32);
: dim3 dimGrid(2, 2);
: 這個配置是以下附的程式碼配置,執行結果是正常的。
: (2)dim3 dimBlock(4, 128);
: dim3 dimGrid(1, 8);
: 換成這樣配置編譯之後,結果卻只有大約三分之一是正常的值,其餘卻都是0
: ,我的threadIdx.x & threadIdx.y都是由0-1023,請問這是發生了什麼錯誤嗎?
: 另外,我想知道執行運算時thread做了什麼事情,每個thread裡面裝的是什麼東西呢,是
: 我輸入的資料嗎?還是加跟乘這兩個指令呢?
: 餵入的資料(Input):
:
: ...
:
: 補充說明(Supplement):
: 請版上的各位先進指導一下我,謝謝。另外,手機排版請見諒。
原文的code恕刪
小弟剛好也在玩cuda,所以來試試看是不是能幫助到原Po,版上高手如雲,還請多指教。
硬體部分的問題恕在下駑頓...(被巴頭,不過你說的顯卡上總共有幾個core,
同時就可以有幾個執行緒在執行這件事情是對的,不過要看你的 Grid, Block 參數
怎麼下,這個部分你必須先從你的顯示卡的計算能力開始了解起,關鍵字:Warp。
如果參數下歪了,可能就會有一小段時間會有幾個 core 在閒置,當然每台機器
最適合的參數都不一樣,我自己也是還在摸索這個部分,這是我目前理解到的程度,
,如果有錯誤的地方還請高手指點。
再來軟體方面我就熟悉多了,cuda真的是很困難debug,在下常常被indexing搞的
不要不要的。首先呢,你說你的 threadIdx.x, threadIdx.y 都是從 0 - 1023,這是
一個大問題,先假設你沒有筆誤,讓我們來算算看,一個 Block 總共的 thread 數目
最多是 1024 (這個部分我不確定你的機器是否一樣,所以你最好先了解一下),好
一般可以把一個 Block 想像成一個三維陣列的 thread,所以一個 Block 總共會有的
thread 數目是 x * y * z (x, y, z)各代表一個維度的執行緒總數,我們把你的
例子套進來,1024 * 1024 * 1,阿搭,這不是就爆了嗎。所以你要確定你的 threadIdx
的 range到底是多少到多少。
再來你說你的輸出有些值對,有些值不對,哈哈哈~哈威(再度被巴,這個問題
有好幾種可能,第一,可能你從頭到尾所有的 thread 只改到 output array 的其中
一部分,所以剩下的部份當然就是0啦,第二種可能,你某些的 Block access 了他
可以用的 memory 但是有些 Block 卻沒有(以為我又要唱歌了嗎 威。這個狀況下
除了你從 code 一行一行去檢查之外,還有一個不錯用的方式,不過我很好奇,既然
你都會用 cudaEvent_t 了,怎麼不會用這個: cudaError_t,這位大大就是你找錯
的好夥伴,實作部分如下:
cudaError_t status;
MyKernel<<< grd, blk >>>( ... ); // 你的名字(X 你的kernel(O
status = cudaGetLastError(); // 問問 cuda 大大我剛剛有沒有做錯什麼
if( status != cudaSuccess ){ // 如果我有做錯 請告訴我錯在哪
cout << cudaGetErrorName( status ) << ":" << cudaGetErrorString( status );
// exit(255); // 看個人決定要不要繼續執行啦
}
使用上大致就是這樣,我個人是會省去 cudaGetErrorName() 的部份,因為我比較懶惰,
原Po可以自己去玩玩看,在決定要怎麼使用,你有你自己的style。
這應該是第三個問題吧,想知道每個 thread 做了什麼,建議妳自己玩玩看這個,
寫一個 kernel 處理一個大小為 1024 的 array,然後只使用一個 Block 來玩,參數
就可以這樣下 <<< 1, blk >>>, blk 可以是 1 - 1024 任何的數字。然後在裡面讓一
個 thread 只改一個 array element,array index 依據 threadIdx.x。這樣你要他
做什麼都可以,最後把它輸出出來自己看看就知道啦。劇透一下每個 thread 其實都
會執行你 kernel 的內容,不同的地方只在於 threadIdx, blockIdx 這些變數的值會
不一樣。
回答的有點簡略,不過希望能夠幫到你。
作者: sunneo (艾斯寇德)   2016-12-04 21:00:00
那只是在表示x,y,z個別的最大值,但仍得保持x*y*z<1024
作者: opl164 (opl)   2016-12-04 20:41:00
一個block內的thread最多是1024個 可是max dimension ofa thread block是(1024,1024,64) 為什麼會這樣?

Links booklink

Contact Us: admin [ a t ] ucptt.com