????CPU主要進(jìn)行串行運(yùn)算,GPU主要用于并行運(yùn)算。GPU運(yùn)算本身就比CPU運(yùn)算快到幾十倍甚至上百倍(和處理器型號以及編程語言有關(guān)),而并行計算更是加快了計算機(jī)的計算速度。下面我用一個很簡單的例子,矢量求和來講解下這個知識點(diǎn)。
????首先來看下基于CPU的矢量求和問題,代碼如下:
????int add(int *a, int *b,int N) {
????int sum = 0;
????for (int i = 0; i < N; i++)
????{
????????sum = sum + a[i] + b[i];
????}
????return sum;
????}
????int main() {
????????int N = 5;
????????int sum = 0;
????????int a[5] = { 1,2,3,4,5 };
????????int b[5] = { 1,2,3,4,5 };
????????sum = add(a, b, N);
????????printf("Twos arrays sum is : %d\n",sum);
????????return 0;
????}
????以上不用做過多解釋,我想大家也都明白,就是簡單的C語言。值得注意的是這就是進(jìn)行的串行運(yùn)算,當(dāng)前一步執(zhí)行完畢后再去執(zhí)行后面的步驟。當(dāng)然CPU也能執(zhí)行并行運(yùn)算,不過要添加一定的代碼創(chuàng)建線程,在這里就不過多講解,有興趣的可以自己找找資料。我們拋磚引玉引出下面的基于GPU的矢量求和,看看GPU到底如何進(jìn)行并行計算的。
????//GPU實(shí)現(xiàn)并行計算
????#define N 10
????__global__ void add(int *a, int *b, int *c) {
????int tid = blockIdx.x; //計算該索引處的數(shù)據(jù)
????????if (tid < N)
????????c[tid] = a[tid] + b[tid];
????}
????int main() {
????????int a[N], b[N], c[N];
????????int *dev_a, *dev_b, *dev_c;
????????//在GPU分配內(nèi)存
????????cudaMalloc((void**)&dev_a, N * sizeof(int));
????????cudaMalloc((void**)&dev_b, N * sizeof(int));
????????cudaMalloc((void**)&dev_c, N * sizeof(int));
????????//CPU上為數(shù)組賦值
????????for (int i = 0; i < N; i++)
????????{
????????????a[i] = -i;
????????????b[i] = i * i;
????????}
????????//將數(shù)組a和b復(fù)制到GPU
????????cudaMemcpy(dev_a, a, N * sizeof(int), cudaMemcpyHostToDevice);
????????cudaMemcpy(dev_b, b, N * sizeof(int), cudaMemcpyHostToDevice);
????????add <<<N, 1 >>> (dev_a, dev_b, dev_c);
????????//將數(shù)組c從GPU復(fù)制到CPU
????????cudaMemcpy(c, dev_c, N * sizeof(int), cudaMemcpyDeviceToHost);
????????//顯示結(jié)果
????????for (int i = 0; i < N; i++) {
? ? ? ? ? ? ?printf("%d + %d = %d\n", a[i], b[i], c[i]);
????????}
????????//釋放GPU上分配的內(nèi)存
????????cudaFree(dev_a);
????????cudaFree(dev_b);
????????cudaFree(dev_c);
????????return 0;
????}
????經(jīng)過觀察發(fā)現(xiàn)上面的代碼我們并不陌生,都是前面講過的,這樣就容易很多了。
????在GPU中運(yùn)算首先肯定要在GPU中開辟內(nèi)存空間,只不過我們需要往dev_a和dev_b中傳入我們在CPU中賦值的需要計算的數(shù)據(jù)。
????前面也說過,GPU運(yùn)算結(jié)束后要進(jìn)行GPU內(nèi)存的釋放,防止造成內(nèi)存泄漏,同CPU原理一樣。
????然后我們需要使用cudaMemcpy()函數(shù)將需要計算的數(shù)據(jù)復(fù)制到設(shè)備中參數(shù)cudaMemcpyHpstToDevice,以及將計算得到的結(jié)果復(fù)制回主機(jī),參數(shù)cudaMemcpyDeviceToHost。其實(shí)如果在GPU上賦值速度會更快,不過我們只是簡單的矢量求和就 不用那么麻煩了。
????當(dāng)我們繼續(xù)閱讀代碼時會看到核函數(shù)的調(diào)用尖括號中的值是N,add<<>>(dev_a,dec_b,dev_c);N是表示設(shè)備在執(zhí)行核函數(shù)時使用的并行線程塊的數(shù)量?,F(xiàn)在是對長度為10的矢量進(jìn)行相加,如果編寫更大規(guī)模的并行應(yīng)用程序,就要將宏定義中的數(shù)字進(jìn)行更改,但是要注意最大值不能超過65535。
????blockIdx變量不需要我們自己定義,該變量是CUDA的一個內(nèi)置變量,表示執(zhí)行設(shè)備代碼的線程塊的索引值。因?yàn)镃UDA支持二維的線程塊數(shù)組,因此它的使用方法是blockIdx.x,將該值賦值給tid,要判斷tid是否小于N,因?yàn)橥ǔG闆r下tid總是小于N,這是在核函數(shù)中這樣假設(shè)的。
????Ok,寫的蠻多的,就這樣。