CUDA

CUDA（Compute Unified Device Architecture，统一计算架构）是由NVIDIA所推出的一種整合技術，是該公司對於GPGPU的正式名稱。透過這個技術，使用者可利用NVIDIA的GeForce 8以後的GPU和較新的Quadro GPU进行计算。亦是首次可以利用GPU作為C-编译器的开发环境。NVIDIA行銷的時候，往往將编译器與架构混合推廣，造成混亂。實際上，CUDA架构可以相容OpenCL或者自家的C-编译器。無論是CUDA C-語言或是OpenCL，指令最終都會被驅動程式轉換成PTX代碼，交由顯示核心計算。

概要

Example of CUDA processing flow
1. 複製記憶體至 GPU
2. CPU 指令驅動 GPU
3. GPU 每一核心并行處理
4. GPU 將結果傳回主記憶體

以GeForce 8800 GTX为例，其核心擁有128个内处理器。利用CUDA技術，就可以將那些内处理器串通起來，成為线程处理器去解决数据密集的计算。而各個内处理器能够交换、同步和共享数据。利用NVIDIA的C-编译器，通過驱动程式，就能利用这些功能。亦能成為流处理器，讓应用程式利用進行運算。

GeForce 8800 GTX显示卡的运算能力可达到520GFlops，如果建設SLI系统，就可以达到1TFlops。

但程序员在利用CUDA技術時，須分開三种不同的存储器，要面對繁复的线程层次，编译器亦无法自动完成多数任务，以上問題就提高了开发难度。而將來的G100會採用第二代的CUDA技術，提高效率，降低开发难度。

目前，已有軟體廠商利用CUDA技術，研發了一個Adobe Premiere Pro的插件。通過插件，使用者就可以利用顯示核心去加速H.264/MPEG-4 AVC的编码速度。速度是單純利用CPU作軟體加速的7倍左右。

在NVIDIA收購AGEIA後，NVIDIA取得相關的物理加速技術，即是PhysX物理引擎。配合CUDA技術，顯示卡可以模擬成一顆PhysX物理加速晶片。目前，全系列的GeForce 8顯示核心都支援CUDA。而NVIDIA亦不會再推出任何的物理加速卡，顯示卡將會取代相關產品。

為了將CUDA推向民用，NVIDIA會舉行一系列的編程比賽，要求参赛者開發程式，充分利用CUDA的計算潛能。但是，要將GPGPU普及化，還要看微軟能否在Windows作業系統中，提供相關的编程接口。

在2008年8月，NVIDIA推出CUDA 2.0。2010年3月22日，NVIDIA推出CUDA 3.0，僅支援Fermi及之後的架構。

應用

利用CUDA技術，配合適當的軟體(例如MediaCoder、Freemake Video Converter)，就可以利用顯示核心進行高清视频編碼加速。视频解碼方面，同樣可以利用CUDA技術實現。此前，NVIDIA的顯示核心本身已集成PureVideo單元。可是，實現相關加速功能的一個微軟API－DXVA，偶爾會有加速失效問題。所以利用CoreAVC配合CUDA，變相在顯示核心上實現軟體解碼，解決兼容性問題。另外，配合適當的引擎，顯示核心就可以計算光线跟踪。NVIDIA就放出了自家的Optix实时光线跟踪引擎，透過CUDA技術利用GPU計算光线跟踪。

支援的產品

所有基於G80及之後架構的民用與專業顯示卡或運算模組皆支援CUDA技術。

Example

下列的範例是如何用 C++ 自GPU的image 陣列中取得紋理（texture）:

cudaArray* cu_array;
texture<float, 2> tex;
 
// Allocate array
cudaChannelFormatDesc description = cudaCreateChannelDesc<float>();
cudaMallocArray(&cu_array, &description, width, height);
 
// Copy image data to array
cudaMemcpy(cu_array, image, width*height*sizeof(float), cudaMemcpyHostToDevice);
 
// Bind the array to the texture
cudaBindTextureToArray(tex, cu_array);
 
// Run kernel
dim3 blockDim(16, 16, 1);
dim3 gridDim(width / blockDim.x, height / blockDim.y, 1);
kernel<<< gridDim, blockDim, 0 >>>(d_odata, height, width);
cudaUnbindTexture(tex);
 
__global__ void kernel(float* odata, int height, int width)
{
   unsigned int x = blockIdx.x*blockDim.x + threadIdx.x;
   unsigned int y = blockIdx.y*blockDim.y + threadIdx.y;
   float c = tex2D(tex, x, y);
   odata[y*width+x] = c;
}

下列的例子是用 Python 改寫. Python 相關的訊息可取自 PyCUDA.

import pycuda.driver as drv
import numpy
import pycuda.autoinit
 
mod = drv.SourceModule("""
__global__ void multiply_them(float *dest, float *a, float *b)
{
  const int i = threadIdx.x;
  dest[i] = a[i] * b[i];
}
""")
 
multiply_them = mod.get_function("multiply_them")
 
a = numpy.random.randn(400).astype(numpy.float32)
b = numpy.random.randn(400).astype(numpy.float32)
 
dest = numpy.zeros_like(a)
multiply_them(
        drv.Out(dest), drv.In(a), drv.In(b),
        block=(400,1,1))
 
print dest-a*b

更多的Python 的矩陣相乘問題可取自 pycublas.

import numpy
from pycublas import CUBLASMatrix
A = 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*B
print C.np_mat()

相關條目

• NVIDIA顯示核心列表

外部連結

• NVIDIA - CUDA（英文）
• NVIDIA - CUDA ZONE（正体中文）
• 周秉誼. GPU高效能運算環境—CUDA與GPU Cluster介紹. 台灣大學計算機及資訊網路中心. [2010-12-05查閱] <span style="font-family: sans-serif; cursor: default; color:#555; font-size: 0.8em; bottom: 0.1em; font-weight: bold;" title="连接到（正体中文）网页">（（正体中文））.