CUDA

CUDA处理流例子
1.将主存的处理数据复制入显存中
2. CPU指令驅動GPU
3. GPU每一核心并行處理
4. GPU將显存結果傳回主記憶體

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

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

但程序员在利用CUDA技術時，須分開三种不同的暫存器，要面對繁复的线程层次，编译器亦无法自动完成多数任务，以上問題就会提高开发难度。而目前更新的G100以及更高核心型号已採用第二代的CUDA技術，可提高效率，降低开发难度。

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

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

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

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

CUDA是一種由NVIDIA提出的並由其製造的圖形處理單元（GPUs）實現的一種平行計算平臺及程式設計模型。CUDA給程式開發人員提供直接訪問CUDA GPUs中的虛擬指令集和平行計算元件的記憶體。

使用CUDA技術，GPUs可以用來進行通用處理（不僅僅是圖形）；這種方法被稱為GPGPU。與CPUs不同的是，GPUs有著側重以較慢速度執行大量併發執行緒的並行流架構，而非快速執行單一執行緒。

軟體發展者可以通過CUDA加速庫，編譯器指令（如OpenACC）以及符合工業標準的程式設計語言（如C,C++和Fortran）擴展對CUDA平臺進行操作。C/C++程式師可以使用“CUDA C/C++”，使用“NVCC”——NVIDIA基於LLVM的C/C++編譯器進行編譯；Fortran程式師可以使用“CUDA Fortran”，使用PGI公司的PGI CUDA Fortran編譯器進行編譯。除了庫、編譯器指令、CUDA C/C++和CUDA Fortran，CUDA平臺還支援其它計算介面，如Khronos Group的OpenCL，Microsoft的DirectCompute，以及C++AMP。其協力廠商封裝也可用於Python，Perl，Fortran，Java，Ruby，Lua，Haskell，MATLAB，IDL及Mathematica的原生支持。

在電腦遊戲行業中，GPUs不僅用於進行圖形渲染，而且用於遊戲物理運算（物理效果如碎片、煙、火、流體），比如PhysX和Bullet。在計算生物學與密碼學等領域的非圖形應用上，CUDA的加速效果達到可以用數量級來表示的程度。

CUDA同時提供底層API與高階API。最初的CUDA軟體發展包（SDK）於2007年2月15日公佈，支持Microsoft Windows和Linux。而後在第二版中加入對Mac OS X的支持，取代2008年2月14日發佈的測試版。所有G8x系列及以後的NVIDIA GPUs皆支援CUDA技術，包括GeForce，Quadro和Tesla系列。CUDA與大多數標準作業系統相容。Nvidia聲明：根據二進位相容性，基於G8x系列開發的程式無需修改即可在未來所有的Nvidia顯卡上運行。

在GPUs（GPGPU）上使用圖形APIs進行傳統通用計算，CUDA技術有下列幾個優點：

• 分散讀取——代碼可以從記憶體的任意位址讀取

• 統一虛擬記憶體（CUDA 4）

• 共用記憶體——CUDA公開一個快速的共用存儲區域（每個處理器48K），使之在多個進程之間共用。其作為一個用戶管理的快取記憶體，比使用紋理查找可以得到更大的有效頻寬。

• 與GPU之間更快的下載與回讀

• 全面支持整型與位操作，包括整型紋理查找

• CUDA不支援完整的C語言標準。它在C++編譯器上運行主機代碼時，會使一些在C中合法（但在C++中不合法）的代碼無法編譯。

• 不支持紋理渲染（CUDA 3.2及以後版本通過在CUDA陣列中引入“表面寫操作”——底層的不透明資料結構——來進行處理）

• 受系統主線的頻寬和延遲的影響，主機與設備記憶體之間資料複製可能會導致性能下降（通過過GPU的DMA引擎處理，非同步記憶體傳輸可在一定範圍內緩解此現象）

• 當執行緒總數為數千時，執行緒應按至少32個一組來運行才能獲得最佳效果。如果每組中的32個進程使用相同的執行路徑，則程式分支不會顯著影響效果；在處理本質上不同的任務時，SIMD執行模型將成為一個瓶頸（如在光線追蹤演算法中遍歷一個空間分割的資料結構）

• 與OpenCL不同，只有NVIDIA的GPUs支援CUDA技術

• 由於編譯器需要使用優化技術來利用有限的資源，即使合法的C/C++有時候也會被標記並中止編譯

• CUDA（計算能力1.x）使用一個不包含遞迴、函數指標的C語言子集，外加一些簡單的擴展。而單個進程必須運行在多個不相交的記憶體空間上，這與其它C語言運行環境不同。

• CUDA（計算能力2.x）允許C++類功能的子集，如成員函數可以不是虛擬的（這個限制將在以後的某個版本中移除）[參見《CUDA C程式設計指南3.1》－附錄D.6]

• 雙精度浮點（CUDA計算能力1.3及以上）與IEEE754標準有所差異：倒數、除法、平方根僅支持舍入到最近的偶數。單精確度中不支持反常值（denormal）及sNaN（signaling NaN）；只支援兩種IEEE舍入模式（舍位與舍入到最近的偶數），這些在每條指令的基礎上指定，而非控制字碼；除法/平方根的精度比單精確度略低。

'*' – 仅限 OEM 产品

详情请参见Nvidia页面存档备份，存于:

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

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

下列的範例是如何用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相關的訊息可取自PyCUDAArchived 2012-03-03 at WebCite.

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()

• OpenCL
• Vulkan

• 官方网站
• 周秉誼. . 台灣大學計算機及資訊網路中心. [2010-12-05]. （原始内容存档于2010-05-16） （中文（繁體）‎）.