CUDA 暫存器和區域性儲存器

2021-10-10 09:44:59 字數 993 閱讀 5803

gpu片上快取記憶體器,基本單元時暫存器檔案,每個暫存器檔案大小為32bit。

計算能力1.0/1.1版本硬體中,每個sm中暫存器檔案數量為8192;而在1.2/1.3硬體中,每個sm中暫存器檔案數量為16384.

一般情況下,核心執行緒中的簡單區域性變數都在暫存器儲存器中。

對於每個執行緒,區域性儲存器也是私有的,如果暫存器被消耗完,資料將被儲存在區域性儲存其中。如果每個執行緒使用了過多的暫存器,或宣告了大型結構體或陣列,或者編譯器無法確定陣列的大小,執行緒的私有資料就有可能會被分配到local memory中。乙個執行緒的輸入和中間變數將被儲存在暫存器或者區域性儲存器中。區域性儲存器中的資料被數儲存在視訊記憶體中,因此對local memory的訪問速度很慢。

如下,mt會被存入local memory中

__global__ void localmemdemo(float* a)

如果在定義執行緒私有陣列的同時,對其就進行了初始化,那麼如果陣列尺寸不大,這個陣列仍然有可能被分到暫存器中。

__global__ void localmemdemo(float* a)

;}

在編譯時,輸出ptx(parallel thread execution)彙編**(在編譯時加上-ptx或者-keep選項),就能觀察變數在編譯的第一階段是否被分配到了local memory中,如果乙個變數在ptx中以.local助記符宣告,可使用ld.local和st.local助記符訪問,這個變數就被存放在local memory中。不過,即使初次編譯的變數不位於local memory中,在編譯的第二階段仍然有可能根據目標硬體中儲存器的大小將變數存放在local memory中。這時,可以通過加上–ptxas-options=-v編譯選項用來觀察lmem的使用情況。

如果陣列較小,並且一定要分配在暫存器中,用以下方法:

__global__ void localmemdemo(float* a)

暫存器 記憶體 儲存器

暫存器 又稱快取 一般是指由基本的rs 觸發器結構衍生出來的 d觸發,就是一些與非門構成的結構,一般整合在 cpu內,其讀寫速度跟 cpu的執行速度基本匹配,但因為效能優越,所以造價昂貴,一般好的 cpu也就只有幾mb的 2級快取,1級快取更小。使用暫存器可以縮短至零長度 節省儲存空間,提高指令的執...

暫存器(通用暫存器)

因為學習使用的是王爽的 組合語言 第3版 因此也只能提到8086cpo的暫存器。對於其他而言,原理都是相通的。對於8086暫存器,有14個暫存器,主要是 ax bx cx dx si di sp bp ip cs ss ds es psw。一 通用暫存器 8086的通用暫存器有ax bx cx dx...

暫存器,標誌暫存器

涉及硬體傳輸資料的,往往包含以下內容 1.資料傳輸引腳 資料放 2.控制引腳 怎麼控制 3.狀態引腳 結果如何 暫存器相當於cpu內部的儲存單元,可能是連續排列,相當於 c語言中的陣列。一 8個通用暫存器 16 bit accumulate axah alr0 count cxch clr1 dat...