nVidia CUDA API(下)

在《nVidia CUDA API(上)》的部分,已經大致說明了 extension 的部分;這邊要來講的則是 runtime library 的部分。

CUDA 的 runtime library 分成下面三部分:

  • common component
    提供 CUDA 的 vector 等型別,以及可以同時在 host 及 device 上可以執行的函式。主要包括:
    • 內建的 vector 型別
    • dim3 型別(用於指定 kernel 參數的型別)
    • texture 型別
    • 數學函式
  • device component
    在 device 上執行的部分,提供 device 上特殊的函式。主要有:
    • 數學函式。這邊的數學函式是 device 的特殊版本,精確度較低,但是速度較快。
    • 同步函式
    • Atomic 函式
    • 型別轉換函式(Conversion/Casting)
    • texture 函式
  • host component
    在 host 上執行的部分,負責控制 device。主要是處理:
    • Device 管理
    • Context 管理(所謂 Context 大致相當於 CPU?)
    • 記憶體管理
    • Code module 管理(Code module 似乎相當於 dynamic library)
    • Execution control
    • Texture reference 管理
    • Interoperability with OpenGL and Direct3D.

不過,這邊應該就不會列舉了∼個別的說明,請參考《CUDA Programming Guide 1.0》這份文件。而這邊,就只大概介紹一些 Heresy 覺得比較會用到的部分了。

記憶體管理

要說最重要、一定會用到的部分,應該就是記憶體管理的部分了!因為在 CUDA 中,要在 device 的程式中存取的資料,一定要先存在 device 上,所以在記憶體管理的部分,CUDA 提供了很類似 C 的一些函式:cudaMalloc()cudaFree()。原則上,cudaMalloc 大致上就是 C 語言中的 malloc,也就像是 C 的 new;而 cudaFree 則就相當於 C 的 free,以及 C 的 delete

而要使用的方法,則是要先宣告 pointer,在 allocate 記憶體給他;下面就是一個簡單的範例:

float	*dev_Array; cudaMalloc( (void**)&dev_Array, ArraySize * sizeof( float ) );

這樣,就可以在 device 上配置一塊大小是 ArraySize 的 float 陣列了∼

如果要把已經存在一般程式中的資料複製到 device 上呢?這時候就要用對應到 C 語言中 memcpy 的函式-cudaMemcpy 了∼以 CUDA SDK 範例裡的寫法,會是:

// define, allocate on host
float *host_Array = new float[100];
for( int i = 0; i < 100; i )
host_Array[i] = i;

// define, allocate on device
float *dev_Array;
cudaMalloc( (void**)&dev_Array, 100 * sizeof( float ) );

// copy from host to device
cudaMemcpy>( dev_Array, host_Array, 100 * sizeof( float ),
cudaMemcpyHostToDevice );

cudaMemcpy 這個函式的四個參數依序為:目標、來源、大小、方法。最後一項方法的型別是 CUDA 的一個列舉型別 cudaMemcpyKind,他有四種值:cudaMemcpyHostToHost, cudaMemcpyHostToDevice, cudaMemcpyDeviceToHost, cudaMemcpyDeviceToDevice;相當直覺的,就是代表由 host/device 複製到 device/host。所以,如果是要反過來把 device 上的資料複製回 host 的話,就只要把最後一項參數改成 cudaMemcpyDeviceToHost 就可以了!

而要釋放掉 allocate 出來的記憶體空間的話,就直接使用 cudaFree 就可以了。

cudaFree( dev_Array );

實際上 CUDA 的記憶體管理除了這邊提到的基本類型,還有 cudaMallocPitch(), cudaMemset(), cudaMemcpy2D(), cudaMemcpyToArray(), cudaMemcpyToSymbol() 等其他的記憶體管理函式,在這邊就不一一提出來了。

裝置管理

一般情況下,可能用不到這一部分的 API 吧∼但是如果有兩個以上的 CUDA Device 的話,就須要透過這些函式來選擇要用的裝置了∼此外,也可以先透過這些函式,來確認機器上的一些基本規格。

這部分的函式不多,有下面幾個:

  • cudaGetDeviceCount( int* count )

    取得目前機器上,支援 CUDA 的 device 數目,會把結果存在 count 中。函式則會 return 一個 cudaError_t,用來表示 CUDA 的錯誤。

  • cudaGetDeviceProperties( cudaDeviceProp* prop, int dev )

    取得目前機器上第 dev 個 device 的屬性;prop 的結構會是:

    struct cudaDeviceProp
    {
    char name[256]; //裝置名稱
    size_t totalGlobalMem; //Device 的 global memory 總量
    size_t sharedMemPerBlock; //Device 的 shared memory 總量
    int regsPerBlock; //每一個 block 的 registers 數量
    int warpSize; //warp size
    size_t memPitch; //最大的 memory pitch
    int maxThreadsPerBlock; //Block 的最大 thread 數
    int maxThreadsDim[3]; //Block 中 thread 最大的維度
    int maxGridSize[3]; //Grid 中 block 的最大維度
    size_t totalConstMem; //Device 的 constant memory 總量
    int major; //主要版本編號
    int minor; //次要版本編號
    int clockRate; //時脈
    size_t textureAlignment; //alignment requirement mentioned
    };
  • cudaChooseDevice( int* dev, const cudaDeviceProp* prop)

    在目前的 device 中,找一個最符合給定的 prop 的,並把他的索引直存在 dev 裡。

  • cudaSetDevice(int dev)

    指定現在要使用第 dev 個 device。

  • cudaGetDevice(int* dev)

    取得目前正在使用的 device 的索引編號,儲存在 dev 中。

快速數學計算

從上面的列表可以發現,在 common component 和 device component 中,都有數學函式的功能。其中,common component 中的應該算是一般的數學計算函式,而 device component 所提供的,則是精確度較低,但是速度較快的版本!

也就是說 CUDA 在 device component 的部分,提供了部分數學計算的函式,是以 GPU 來做計算的快速版本。這些函式在 CUDA 中統一以「__」來做開頭,例如 __sinf(x)__fadd_rz(x,y)__logf(x) 等等。這些函式在《CUDA Programming Guide 1.0》的 Appendix B 中,可以找到比較完整的列表可以參考。

而在 nvcc 的編譯參數中,也還提供了一個「-use_fast_math」的參數,可以強制所有可以的數學計算函數,都使用叫快速的版本來計算。這在精確度比較不重要的情況下,應該可以對效能有些增益。


上面講的部分,都是 runtime API 的部分;大部分的函式,都是以 cuda 來做為開頭。而實際上,CUDA 還有提供 Driver API;這一部分的函式會是以 cu 來做為開頭。而許多函式,都可以和 runtime API 來做一一對應。而以文件裡的說法,Driver API 應該是比 runtime API 來的低階,但是實際的差異,Heresy 也不是很清楚就是了。


本文主要參考nVidia 的《CUDA Programming Guide 1.0》(PDF 文件)的第四章。


原始發表: nVidia CUDA API(下)

5 thoughts on “nVidia CUDA API(下)”

  1. 錯字:可以強制所有可以的數學計算函數,都使用 “叫” 快速的版本來計算。

  2. 你好,常靠你的幫忙進行kinect的開發,主要是利用她去抓物體的3D位置,傳給另一個程式,去做遠端遙控機器手臂,想請問的是,像這樣的開發,除了裝好openNI VS2010,還需要向openCV來把得到的影像及深度資料繪成圖嗎? 還是直接用openNI就行呢?(我只要能將物件用滑鼠標出抓出3D位置傳給另一個程式就好)拜託版主幫忙解答了QQ

  3. to YUCK 麻煩請把問題發表於對應的文章,這樣會比較一致,謝謝。另外,這部分的重點應該是看你兩個程式之間要怎麼溝通,感覺和 OpenCV 似乎沒什麼關係?除非你是打算把它存成圖檔讓另一個程式來讀取。

  4. 恩恩,抱歉因為PO了才發現弄錯版 謝謝版主來回答。

發佈留言

發佈留言必須填寫的電子郵件地址不會公開。 必填欄位標示為 *