CUDA

CUDA仍依賴於外部的編譯器，如Microsoft Windows下需要Microsoft Visual Studio；Linux下則主要依賴於GCC。簡介.GPU不僅用於進行圖形渲染，而且用於物理運算如PhysX和Bullet。進一步的，GPU可以用在計算生物學與密碼學等領域的非圖形應用上。在NVIDIA收購AGEIA後，NVIDIA取得相關的物理加速技術，即是PhysX物理引擎。配合CUDA技術，顯示卡可以模擬成一顆PhysX物理加速晶片。目前，全系列的GeForce 8顯示核心都支援CUDA。而NVIDIA亦不會再推出任何的物理加速卡，顯示卡將會取代相關產品。而使用CUDA技術，GPU可以用來進行通用處理；這種方法被稱為GPGPU。與CPU不同的是，GPU以較慢速度並行大量執行緒，而非快速執行單一執行緒。以GeForce 8800 GTX為例，其核心擁有128個內處理器。利用CUDA技術，就可以將那些內處理器做為執行緒處理器，以解決資料密集的計算。而各個內處理器能夠交換、同步和共享資料。GeForce 8800 GTX的運算能力可達到520GFlops，如果建設SLI系統，就可以達到1TFlops。目前，已有軟體廠商利用CUDA技術，研發出Adobe Premiere Pro的外掛。透過外掛，使用者就可以利用顯示核心去加速H.264/MPEG-4 AVC的編碼速度。速度是單純利用CPU作軟體加速的7倍左右。雖然CUDA底層是以C/C++為主，並以使用“NVCC”——NVIDIA基於LLVM的C/C++編譯器介面來進行編譯，但程式師也可以使用編譯器指令以及多種程式設計語言擴充套件對CUDA平臺進行操作。如Fortran程式師可以使用“CUDA Fortran”，或PGI公司的PGI CUDA Fortran 編譯器進行編譯。除了此之外CUDA平臺還支援其它計算介面，如Khronos Group的OpenCL，Microsoft的DirectCompute，以及C++AMP。也可以透過其他語言如 Python，Perl，Java，Ruby，Lua，Haskell，MATLAB，IDL及Mathematica 的介面間接呼叫CUDA。CUDA最初的CUDA軟體發展包於2007年2月15日公佈，同時支援Microsoft Windows和Linux。而後在第二版中加入對Mac OS X的支援(但於CUDA Toolkit 10.2起放棄對macOS的支援)，取代2008年2月14日釋出的測試版。所有G8x系列及以後的NVIDIA GPUs皆支援CUDA技術，包括GeForce，Quadro和Tesla系列。CUDA與大多數標準作業系統相容。Nvidia宣告：根據二進位相容性，基於G8x系列開發的程式無需修改即可在未來所有的Nvidia顯示卡上執行。優點.在GPUs上使用圖形APIs進行傳統通用計算，CUDA技術有下列幾個優點：顯示卡的受支援情況.'*' – 僅限 OEM 產品詳情請vidia:應用.利用CUDA技術，配合適當的軟體，就可以利用顯示核心進行高畫質影片編碼加速。影片解碼方面，同樣可以利用CUDA技術實現。此前，NVIDIA的顯示核心本身已整合PureVideo單元。可是，實現相關加速功能的一個微軟API－DXVA，偶爾會有加速失效問題。所以利用CoreAVC配合CUDA，變相在顯示核心上實現軟體解碼，解決相容性問題。另外，配合適當的引擎，顯示核心就可以計算光線跟蹤。NVIDIA就放出了自家的Optix實時光線跟蹤引擎，透過CUDA技術利用GPU計算光線跟蹤。支援的產品.所有基於G80及之後架構的民用與專業顯示卡或運算模組皆支援CUDA技術。示例.CUDA Driver API.下面將示範以最底層的CUDA Driver API 呼叫GPU做列向量的加法，以下為 CPU 端的程式碼// 本範例修改自 Andrei de A. Formiga (2012-06-04) 寫的範例: https://gist.github.com/tautologico/2879581// 編譯指令 nvcc -O3 -lcuda add.c -o add.exe// 利用CUDA函式的錯誤傳回做例外處理inline void checkCudaErrors( CUresult err)if( CUDA_SUCCESS != err) {printf("CUDA Driver API error = %04d from file , line %i.\n",err, __FILE__, __LINE__ );exit(-1); // 直接終止程式CUdevice device; // CUDA 裝置(也就是GPU)物件CUcontext context; // CUDA 內容物件CUmodule module; // 代表GPU程式碼的物件CUfunction function; // CUDA GPU 函式size_t totalGlobalMem; // CUDA 裝置記憶體總量// Driver API 只能自外部檔案讀取 GPU 程式，可以為 PTX 中間碼也可以是 cubin 機器碼(或是混合各種架構機器碼的fatbin)char *module_file = (char*) "matSumKernel.cubin";// GPU 函式名稱char *kernel_name = (char*) "matSum";// 初始化 CUDA 的手續void initCUDA()int deviceCount = 0; // 當前可使用的 CUDA 裝置(GPU)數CUresult err = cuInit(0); // 初始化 CUDA APIif (err == CUDA_SUCCESS) // 取得可用裝置數checkCudaErrors(cuDeviceGetCount(&amp;deviceCount));if (deviceCount == 0) { // 確定有可用的裝置fprintf(stderr, "Error: no devices supporting CUDA\n");exit(-1);// get first CUDA devicecheckCudaErrors(cuDeviceGet(&amp;device, 0)); // 取編號為 0 的裝置char name[100];cuDeviceGetName(name, 100, device); // 印出裝置名稱printf("&gt; Using device 0: %s\n", name);checkCudaErrors( cuDeviceTotalMem(&amp;totalGlobalMem, device) );// 印出裝置可用記憶體printf(" Total amount of global memory: %llu bytes\n",(unsigned long long)totalGlobalMem);// GPU 記憶體是否是為64bits定址printf(" 64-bit Memory Address: %s\n",(totalGlobalMem &gt; (unsigned long long)4*1024*1024*1024L)?"YES" : "NO");// 建立 CUDA 內容err = cuCtxCreate(&amp;context, 0, device);if (err != CUDA_SUCCESS) {fprintf(stderr, "* Error initializing the CUDA context.\n");cuCtxDetach(context);exit(-1);// 讀取編譯好的cubin GPU程式碼err = cuModuleLoad(&amp;module, module_file);if (err != CUDA_SUCCESS) {fprintf(stderr, "* Error loading the module %s\n", module_file);cuCtxDetach(context); // 釋放 CUDA 內容物件exit(-1);// 獲取GPU程式裡函式"matSum"的指標err = cuModuleGetFunction(&amp;function, module, kernel_name);if (err != CUDA_SUCCESS) {fprintf(stderr, "* Error getting kernel function %s\n", kernel_name);cuCtxDetach(context);exit(-1);int main(int argc, char **argv)int a[N], b[N], c[N];CUdeviceptr d_a, d_b, d_c;// 注意 GPU 變數指標的型態是 CUdeviceptr// typedef unsigned int CUdeviceptr_v2// typedef CUdeviceptr_v2 CUdeviceptr// 初始化主記憶體變數for (int i = 0; i &lt; N; ++i) {a[i] = i;b[i] = N - i;initCUDA();// 動態分配 GPU 記憶體// CUresult cuMemAlloc ( CUdeviceptr* dptr, size_t bytesize )checkCudaErrors( cuMemAlloc(&amp;d_a, sizeof(int) * N) ); //checkCudaErrors( cuMemAlloc(&amp;d_b, sizeof(int) * N) );checkCudaErrors( cuMemAlloc(&amp;d_c, sizeof(int) * N) );// 將列向量傳入裝置// CUresult cuMemcpyHtoD ( CUdeviceptr dstDevice, const void* srcHost, size_t ByteCount )checkCudaErrors( cuMemcpyHtoD(d_a, a, sizeof(int) * N) );checkCudaErrors( cuMemcpyHtoD(d_b, b, sizeof(int) * N) );void *args[3] = { &amp;d_a, &amp;d_b, &amp;d_c }; // 包裝放入GPU 函式的引數// 執行 GPU 函式// CUresult cuLaunchKernel ( CUfunction f, unsigned int gridDimX, unsigned int gridDimY, unsigned int gridDimZ,// unsigned int blockDimX, unsigned int blockDimY, unsigned int blockDimZ,// unsigned int sharedMemBytes, CUstream hStream, void** kernelParams, void** extra )checkCudaErrors( cuLaunchKernel(function, N, 1, 1, // Nx1x1 blocks1, 1, 1, // 1x1x1 threads0, 0, args, 0) );// 將運算結果送回主記憶體// CUresult cuMemcpyDtoH ( void* dstHost, CUdeviceptr srcDevice, size_t ByteCount )checkCudaErrors( cuMemcpyDtoH(c, d_c, sizeof(int) * N) );// 將 CPU 和 GPU 運算結果做對照for (int i = 0; i &lt; N; ++i) {if (c[i] != a[i] + b[i])printf("* Error at array position %d: Expected %d, Got %d\n",i, a[i]+b[i], c[i]);// 釋放 GPU 記憶體// CUresult cuMemFree ( CUdeviceptr dptr )checkCudaErrors( cuMemFree(d_a) );checkCudaErrors( cuMemFree(d_b) );checkCudaErrors( cuMemFree(d_c) );cuCtxDetach(context); // 釋放 CUDA 內容物件return 0;而以下是GPU端的程式碼// 本範例修改自 Andrei de A. Formiga (2012-06-04) 寫的範例: https://gist.github.com/tautologico/2879581// 此部分要先編譯成 cubin 後才可以被 CPU 端程式使用// 編譯指令 nvcc -O3 -cubin -arch=native matSumKernel.cu -o matSumKernel.cubinextern "C" __global__ void matSum(int *a, int *b, int *c)int tid = blockIdx.x; // thread 的 x 座標if (tid &lt; N)c[tid] = a[tid] + b[tid]; //每個 thread 做一次加法CUDA Runtime API.下列的範例是以相較於 Driver API 來說比較簡便的 CUDA Runtime API  做列向量的加法：// 本範例修改自Nvidia官方的CUDA開發指引: https://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#kernels// 編譯指令 nvcc vector_add.cu -arch=native -o vector_add.exe// -arch=native 代表將 device code 編譯成當前電腦 Nvidia GPU 架構的機器碼，拿掉就是照預設編譯成 PTX 中間碼。// Device code: 送入GPU執行的部分__global__ void VecAdd(float* A, float* B, float* C)int i = threadIdx.x; // thread 的 x 座標if (i &lt; N){C[i] = A[i] + B[i]; // 每個 thread 作一次加法// Host code: 送入CPU執行的部分int main()size_t size = N * sizeof(float); // 向量的實際大小，以位元組(bytes)為單位int i; // 迴圈計數// 動態分配位於"host(CPU) 記憶體" 的向量float* h_A = (float*)malloc(size);float* h_B = (float*)malloc(size);float* h_C = (float*)malloc(size);// 隨機初始化輸入向量for(i = 0; i &lt; N; i++){h_A[i] = (float)rand() / (float)RAND_MAX;h_B[i] = (float)rand() / (float)RAND_MAX;// 動態分配位於"device(GPU) 記憶體"的向量float* d_A;cudaMalloc(&amp;d_A, size); // cudaError_t cudaMalloc ( void** devPtr, size_t size )float* d_B;cudaMalloc(&amp;d_B, size);float* d_C;cudaMalloc(&amp;d_C, size);// 將向量從 CPU 複製到 GPUcudaMemcpy(d_A, h_A, size, cudaMemcpyHostToDevice);cudaMemcpy(d_B, h_B, size, cudaMemcpyHostToDevice);// 將 device code 送入 GPU 並執行，執行時一個 Grid 只有一個 block ，一個 block 有 N 個 threadVecAdd«(d_A, d_B, d_C);// 將算好的向量從 GPU 複製到 CPUcudaMemcpy(h_C, d_C, size, cudaMemcpyDeviceToHost);// 印出運算結果for(i = 0; i &lt; N; i++){printf("%f ", h_C[i]);// 釋放 GPU 記憶體cudaFree(d_A);cudaFree(d_B);cudaFree(d_C);// 釋放 CPU 記憶體free(h_A);free(h_B);free(h_C);Python.以下以pycuda ，間接呼叫 CUDA 做列向量的阿達瑪乘積(也就是元素間乘積而非內積)import pycuda.driver as drv #CUDA driversimport pycuda.autoinit # 自動初始化CUDAimport numpy # 矩陣運算mod = drv.SourceModule("""__global__ void multiply_them(float *c, float *a, float *b)int i = threadIdx.x;c[i] = a[i] * b[i];""")multiply_them = mod.get_function("multiply_them")h_a = numpy.random.randn (400).astype(numpy.float32)h_b = numpy.random.randn (400).astype(numpy.float32)h_c = numpy.zeros_like (a)執行GPU的函式:注意這裡h_a, h_b 和 h_c 是在CPU記憶體的 python 變數以下按照原來 GPU 程式碼的區域變數順序，指定哪些是從CPU傳入GPU (drv.In)；哪些是從GPU傳入CPU (drv.Out)，簡單來說，自 GPU 傳出到 h_c； h_a 傳入 GPU；h_b 傳入 GPU。(400,1,1) 代表一個block裡有的thread數量為 400 x 1 x 1multiply_them(drv.Out(h_c), drv.In(h_a), drv.In(h_b),block=(400,1,1))print (d_c)也可以用pycublas 間接呼叫 CUDA ，來計算矩陣乘法import numpyfrom pycublas import CUBLASMatrixA = CUBLASMatrix(numpy.mat(1,2,3],[4,5,6,numpy.float32))B = CUBLASMatrix(numpy.mat(2,3],[4,5],[6,7,numpy.float32))C = A*Bprint(C.np_mat())