CUDA下的GPU程式設計 執行緒和變數

2021-06-18 14:19:24 字數 1882 閱讀 5988

cuda的執行緒是多維的,啟動乙個執行緒格,執行緒格可以是多維的,執行緒格中分為執行緒塊,執行緒塊也可以是多維的,執行緒塊中包含執行緒,對於gpu來說,同時啟動200萬個執行緒是很輕易的事情。

blockidx :當前執行裝置**的執行緒塊的索引

dim3 grid(dim,dim):宣告乙個二維的執行緒格,名字為grid;

kernel<<>(dev_t):執行緒塊為乙個二維執行緒塊,每個執行緒塊有乙個執行緒

griddim:儲存執行緒格每一維的執行緒塊的數量

blockdim:儲存的是執行緒塊中每一維的執行緒數量

若執行緒塊和執行緒格都是一維的:

blockdim.x * griddim.x表示的是總共執行緒數量。

cuda定義的函式和變數方式有__global__ ­ __device__ 和__shared__,

__global__函式直接告訴編譯器,這個函式是在gpu上執行的,不是在cpu上執行,可以在主函式中直接呼叫(主函式中的**如果沒有特別說明都是在cpu上執行的),此時該函式會自動交由gpu執行。__global__只能寫在主函式中。

宣告格式:__global__ add(int *a){}

呼叫格式:add<<>>(dev_t);(啟動執行緒塊為n,每個執行緒塊有m執行緒的執行緒集去執行add(dev_t)函式.

__device__可以定義gpu全域性變數和函式,表示該變數或函式是儲存在gpu上的,**也是在gpu上執行的,與__global__不同的是,__device__只能在__device__函式或__global__中呼叫,不能直接寫在主函式。

可以直接在主函式中宣告乙個普通的變數指標,如int *dev_t,然後再由主函式中呼叫handle_error(cudamalloc(  (void**) &dev_t,sizeof(int) ) );就可以給dev_t分配乙個gpu記憶體上的位址,也就是說dev_t指向的是gpu的位址,此後不能在cpu執行的**中呼叫該指標,因為cpu的記憶體跟gpu的記憶體不是共享的,只能在__global__或者是__device__函式中呼叫該指標。如果要實現cpu和gpu的通訊,只能通過拷貝位址函式將gpu上某段記憶體拷貝到主存上。比如若gpu中計算好了乙個值在指標dev_t所指的int區域,要返回給cpu的t變數話,handle_error(cudamemcpy( &t , dev_t, sizeof(int),cudamemcpydevicetohost ));

__shared__宣告的變數,這個變數會駐留在共享記憶體中,對於在gpu上啟動的每個執行緒塊,cuda編譯器都會建立該變數的乙個副本,執行緒塊中的每個執行緒都共享這塊記憶體,但是執行緒卻無法看到也不難修改其他執行緒塊的變數副本。使得乙個執行緒塊中的多個執行緒能在計算機上進行通訊和協作,訪問共享記憶體的延遲要遠低於訪問普通緩衝區的延遲。

__constant__ 常量記憶體儲存在裝置函式執行期間不會發生變化的資料,nvidia硬體提供了64kb的常量記憶體。常量記憶體裡的變數宣告也要加上__constant__修飾符,如果是陣列的話,要在編譯時為這個陣列提交乙個固定的大小,如__constant__ int a[10]。

當從主機記憶體複製到gpu上的常量記憶體時,不使用cudamemcpy(),用cudamemcpytosymbol()。

執行緒束是指乙個包含32個執行緒的集合,多個執行緒集合被「編織在一起」並以步調一致「的形式執行,在程式中的每一行,執行緒束中的每個執行緒都將在不同的資料上執行相同的指令。如果在半執行緒束中的每個執行緒都從常量記憶體的相同位址上讀取資料,那麼gpu只會產生一次讀取請求並在隨後將資料廣播到每個執行緒,如果從常量記憶體中讀取大量的資料,那麼這種方式產生的記憶體流量指示全域性時的1/16。但是當半執行緒束中的所有16個執行緒需要訪問常量記憶體中不同資料,那麼這個16次不同的讀取請求操作會被序列化,反而影響效率。

裝置執行的**中可以使用__syncthreads()進行對執行緒塊中的執行緒進行同步,可以確保執行緒塊中的每個執行緒都執行完__syncthreads()前面的語句後,才會執行下一條語句。

GPU 高效能程式設計 CUDA 執行緒協作

並行執行緒塊的分解 在向量加法中,為向量中的每乙個元素都啟動乙個執行緒塊 add dev a,dev b,dev c 尖括號中的第乙個引數建立的執行緒塊的數量,第二個引數表示每個執行緒塊中建立的執行緒數量,所以上述啟動的執行緒數量為 n n 1 使用執行緒實現 gpu 上的向量求和 需要修改兩個地方...

GPU高效能程式設計CUDA實戰 第5章執行緒協作

add dev a,dev b,dev c 第乙個引數 啟動的執行緒塊數量。第二個引數 每個執行緒塊中建立的執行緒數量。1.使用執行緒實現gpu上的向量求和 add loop blocks.cu 啟動n個執行緒塊,每個執行緒塊對應乙個執行緒 add dev a,dev b,dev c 改為啟動n個執...

Linux下的多執行緒程式設計

執行緒是作業系統能夠進行排程運算的最小單位,它被包含在程序之中,是程序中的實際運作單位。一條執行緒指的是程序中乙個單一順序的控制流,乙個程序可以併發多個執行緒,每條執行緒執行不同的任務。include intpthread create pthread t pthread,const pthread...