CUDA 版
而如果要寫成 CUDA 的，向量相加的函式會變成：
{
int idx = threadIdx.x;
output[idx] = arrayA[idx] + arrayB[idx];
}
void add_vector_gpu( float* a, float* b, float *c, int size )
{
int data_size = size * sizeof(float);
// part1, allocate data on device
float *dev_A, *dev_B, *dev_C;
cudaMalloc( (void**)&dev_A, data_size );
cudaMalloc( (void**)&dev_B, data_size );
cudaMalloc( (void**)&dev_C, data_size );
// part2, copy memory to device
cudaMemcpy( dev_A, a, data_size, cudaMemcpyHostToDevice );
cudaMemcpy( dev_B, b, data_size, cudaMemcpyHostToDevice );
// part3, run kernel
VectorAdd<<< 1, size >>>( dev_A, dev_B, dev_C );
// part4, copy data from device
cudaMemcpy( c, dev_C, data_size, cudaMemcpyDeviceToHost );
// part5, release data
cudaFree(dev_A);
cudaFree(dev_B);
cudaFree(dev_C);
}
而在 main function 的呼叫，和 CPU 的版本一樣，只要寫：
就可以了～
從上面的程式可以發現，實際在做加法的函式是「VectorAdd」這個函式;在寫的時候，他前面有加上「__global__」來宣告成 CUDA 的 kernel function。而實際用來呼叫的「add_vector_gpu」這個函式，則算是用來把 CUDA 的相關程式封包起來;在裡面除了要讓 device 執行「VectorAdd」這個 GPU 的 kernel 程式外，還要把記憶體由 host(系統記憶體)複製到 device(顯示卡的記憶體)上，也就是 part1, part2 的部分。此外，在「VectorAdd」執行完後，還要再把結果由 device 複製回 host(part4)，並且把 device 上的資源釋放掉。所以其實和 CPU 程式比起來，還是稍微繁瑣了些。
下面大概講一下整個流程：
- part1 的部分，是用來宣告並透過 cudaMalloc 指派記憶體空間在 device 上
- part2 則是透過 cudaMemcpy 把 a, b 的資料從 host 複製到 device
- part3 執行 kernel 程式：VectorAdd。而「<<< 1, size >>>」的部分，就是指定用 1*1 的 grid，裡面則是 size*1*1 的 thread block;也就是說，總共會有一個 grid，grid 裡只有一個 block，block 裡有 size 個 thread。而執行的結果，就相當於 VectorAdd 會被呼叫 size 次，每次的 threadIdx.x 都會是不同的值。
- part4 則是透過 cudaMemcpy 把計算完的 dev_C 的資料從 device 複製回 host
- part5 會將 dev_A, dev_B, dev_C 這些已經用不到的記憶體空間，透過 cudaFree 來釋放掉。
另外要注意的一點，這份程式為了簡單化，在kernel 程式的設定，採用了把所有 thread 放在同一個 block 的設定;在這個情況下，如果 size 比 device 中每個 block 的最大 thread 數還多的話，會沒辦法正確執行。這個時候，就須要把把它拆成幾個 thread block 來執行了～
這份程式，應該就算是 Heresy 對於 CUDA 的 Hellow World 吧～原始碼的話，有放一份在 Microsoft 的 SkyDrive 上(下載：VectorAdd.zip)，有興趣的人可以抓下來試試看。
而由於 Heresy 個人是希望可以把 CUDA 的程式和一般的 C/C++ 程式分開來，所以在 Heresy 的測試專案裡，是分成了 VectorAdd.cpp 和 VectorAdd.cu 兩個檔案。main 是寫在 VectorAdd.cpp 中，而為了要讓 main 裡能夠呼叫 VectorAdd.cu 裡的 add_vector_gpu，所以必須要在兩個檔案都加入
把 add_vector_gpu 這個函式特別標註成是用 C 的方式來編譯。
在這份程式裡，CPU 的程式和 CUDA 的程式都有;在最後的部分，有驗證兩者的結果，理論上應該會是都一樣的～
