在前面的章節中,我們不止一次看到了在呼叫定義的核函式時採用了類似下面的形式:
kernel<<<1,1>>>(param1,param2,...)
「<<< >>>」中引數的作用是告訴我們該如何啟動核函式(比如如何設定執行緒)。 下面我們先直接介紹引數概念,然後詳細說明其意義。
4.1 核函式執行引數
當我們使用gloabl宣告核函式後
__global__ void kernel(param list)
在主機端(host)呼叫時採用如下的形式:
kernel<<>>(param list);
4.2 執行緒結構關於cuda的執行緒結構,有著三個重要的概念:grid, block, thread下圖是乙個結構關係圖:
此外,block, thread的組織結構可以是可以是一維,二維或者三維。以上圖為例,block, thread的結構分別為二維和三維。
cuda中每乙個執行緒都有乙個唯一標識threadidx,這個id隨著組織結構形式的變化而變化。 (注意:id的計算,同計算行優先排列的矩陣元素id思路一樣。)
回顧之前我們的向量加法:
// block是一維的,thread也是一維的
__global__ void addkernel(int *c, const
int *a, const
int *b)
// block是一維的,thread是二維的
__global__ void addkernel(int
*c, int
*a, int
*b)
// block是二維的,thread是三維的
__global__ void addkernel(int *c, int *a, int *b)
下表是不同計算能力的gpu的技術指標(更多可參見 cuda toolkit documentation)
當然也可以通過下面的**來直接查詢自己gpu的具體指標:
#include "cuda_runtime.h"
#include
int main()
system("pause");
return
0;}
其中cudadeviceprop是乙個定義在driver_types.h中的結構體,大家可以自行檢視其定義。
4.3 記憶體結構
如下圖所示,每個 thread 都有自己的乙份 register 和 local memory 的空間。同乙個 block 中的每個 thread 則有共享的乙份 share memory。此外,所有的 thread(包括不同 block 的 thread)都共享乙份 global memory、constant memory、和 texture memory。不同的 grid 則有各自的 global memory、constant memory 和 texture memory。
這種特殊的記憶體結構直接影響著我們的執行緒分配策略,因為需要通盤考慮資源限制及利用率。 這些後續再進行討論。
4.4 異構程式設計
如下圖所示,是常見的gpu程式的處理流程,其實是一種異構程式,即cpu和gpu的協同。
主機上執行序列**,裝置上則執行並行**。
GPU程式設計自學3 CUDA程式初探
3.1 主機與裝置 通常將cpu及其記憶體稱之為主機,gpu及其記憶體稱之為裝置。如下圖所示,新建乙個nvidia cuda工程,並命名為 1 helloworld 之後發現專案裡多了乙個 kernel.cu 的檔案,該檔案內容是乙個經典的向量相加的gpu程式。可以暫時全部注釋該 並嘗試編譯執行下面...
GPU計算與CUDA程式設計
1.來自nvidia的 cuda 並行程式設計框架是 gpgpu 正規化的一種特定的實現。它提供了gpu程式設計的簡易介面,基於cuda程式設計可以構建基於gpu計算的應用程式。2.cuda 在技術上是一種異構計算環境,也就是說它利用了同時在 cpu 和 gpu 上的協調計算。cuda 架構由主機 ...
GPU 高效能程式設計 CUDA 執行緒協作
並行執行緒塊的分解 在向量加法中,為向量中的每乙個元素都啟動乙個執行緒塊 add dev a,dev b,dev c 尖括號中的第乙個引數建立的執行緒塊的數量,第二個引數表示每個執行緒塊中建立的執行緒數量,所以上述啟動的執行緒數量為 n n 1 使用執行緒實現 gpu 上的向量求和 需要修改兩個地方...