CPU主要进行串行运算,GPU主要用于并行运算。GPU运算本身就比CPU运算快到几十倍甚至上百倍(和处理器型号以及编程语言有关),而并行计算更是加快了计算机的计算速度。下面我用一个很简单的例子,矢量求和来讲解下这个知识点。
首先来看下基于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语言。值得注意的是这就是进行的串行运算,当前一步执行完毕后再去执行后面的步骤。当然CPU也能执行并行运算,不过要添加一定的代码创建线程,在这里就不过多讲解,有兴趣的可以自己找找资料。我们抛砖引玉引出下面的基于GPU的矢量求和,看看GPU到底如何进行并行计算的。
//GPU实现并行计算
#define N 10
__global__ void add(int *a, int *b, int *c) {
int tid = blockIdx.x; //计算该索引处的数据
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分配内存
cudaMalloc((void**)&dev_a, N * sizeof(int));
cudaMalloc((void**)&dev_b, N * sizeof(int));
cudaMalloc((void**)&dev_c, N * sizeof(int));
//CPU上为数组赋值
for (int i = 0; i < N; i++)
{
a[i] = -i;
b[i] = i * i;
}
//将数组a和b复制到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);
//将数组c从GPU复制到CPU
cudaMemcpy(c, dev_c, N * sizeof(int), cudaMemcpyDeviceToHost);
//显示结果
for (int i = 0; i < N; i++) {
printf("%d + %d = %d\n", a[i], b[i], c[i]);
}
//释放GPU上分配的内存
cudaFree(dev_a);
cudaFree(dev_b);
cudaFree(dev_c);
return 0;
}
经过观察发现上面的代码我们并不陌生,都是前面讲过的,这样就容易很多了。
在GPU中运算首先肯定要在GPU中开辟内存空间,只不过我们需要往dev_a和dev_b中传入我们在CPU中赋值的需要计算的数据。
前面也说过,GPU运算结束后要进行GPU内存的释放,防止造成内存泄漏,同CPU原理一样。
然后我们需要使用cudaMemcpy()函数将需要计算的数据复制到设备中参数cudaMemcpyHpstToDevice,以及将计算得到的结果复制回主机,参数cudaMemcpyDeviceToHost。其实如果在GPU上赋值速度会更快,不过我们只是简单的矢量求和就 不用那么麻烦了。
当我们继续阅读代码时会看到核函数的调用尖括号中的值是N,add<<>>(dev_a,dec_b,dev_c);N是表示设备在执行核函数时使用的并行线程块的数量。现在是对长度为10的矢量进行相加,如果编写更大规模的并行应用程序,就要将宏定义中的数字进行更改,但是要注意最大值不能超过65535。
blockIdx变量不需要我们自己定义,该变量是CUDA的一个内置变量,表示执行设备代码的线程块的索引值。因为CUDA支持二维的线程块数组,因此它的使用方法是blockIdx.x,将该值赋值给tid,要判断tid是否小于N,因为通常情况下tid总是小于N,这是在核函数中这样假设的。
Ok,写的蛮多的,就这样。