跑一个最简单的Cuda程序

好，现在我们先来看一个最简单的Cuda程序。就像你们学习C语言的时候，直接运行里一个打印“Hello，world”的程序之后老师才会开始跟你解释，我们边练边学。

#include <stdio.h>

__global__ void vector_add(const int *a, const int *b, int *c) {
    *c = *a + *b;
}

int main(void) {
    const int a = 2, b = 5;
    int c = 0;

    int *dev_a, *dev_b, *dev_c;

    cudaMalloc((void **)&dev_a, sizeof(int));
    cudaMalloc((void **)&dev_b, sizeof(int));
    cudaMalloc((void **)&dev_c, sizeof(int));

    cudaMemcpy(dev_a, &a, sizeof(int), cudaMemcpyHostToDevice);
    cudaMemcpy(dev_b, &b, sizeof(int), cudaMemcpyHostToDevice);

    vector_add<<<1, 1>>>(dev_a, dev_b, dev_c);

    cudaMemcpy(&c, dev_c, sizeof(int), cudaMemcpyDeviceToHost);

    printf("%d + %d = %d, Is that right?\n", a, b, c);

    cudaFree(dev_a);
    cudaFree(dev_b);
    cudaFree(dev_c);

    return 0;
}

先用nsight创建一个Cuda运行时项目，然后把代码抄进去，运行一下。

如果你喜欢命令行的方式，你可以把代码抄进一个文本里面，命名为vector_add.cu，然后执行命令nvcc vector_add.cu，然后执行编译生成的a.out文件。

没错，nvcc就像gcc一样用，大多数功能都是相同的，至于.cu这个后缀，这仅仅是个为了说明这是个cuda的程序而已。

如果顺利的话，你应该会得到这样的结果输出

2 ＋ 5 ＝ 7. Is that right?

没错，得到这样的结果就说明的Cuda程序已经可以正确运行了。

现在，让我们来分析一下，上面的代码究竟做了些什么东西。

/*
 * 为了方便起见，我甚至把一些出错判断都去掉了。
 * 比如申请内存空间，谁都没法保证内存一定有空间给你申请的不是吗？
 * 之所以这样，是为了方便我们看清楚一个基本的cuda程序的执行流程。
 */
#include <stdio.h>

/*
 * 这个是要放在显卡上面跑的函数，除了前面多了一个__global__之外，
 * 这个函数其实很简单不是吗？
 */
__global__ void vector_add(const int *a, const int *b, int *c) {
    *c = *a + *b;
}

int main(void) {
    //这里定义了三个整形变量a，b和c。他们位于主存上。
    const int a = 2, b = 5;
    int c = 0;

    //这里定义了三个整形指针。为什么要用指针？
    //因为这是在显存上申请空间，对于主机来说，它只能也只需要知道一个映射的地址就够了。
    int *dev_a, *dev_b, *dev_c;

    //为三个指针申请显存空间，虽然只是一个int，4个字节的大小。
    cudaMalloc((void **)&dev_a, sizeof(int));
    cudaMalloc((void **)&dev_b, sizeof(int));
    cudaMalloc((void **)&dev_c, sizeof(int));

    //将内存上的数据复制到显存上，这里复制了a和b两个加数。
    cudaMemcpy(dev_a, &a, sizeof(int), cudaMemcpyHostToDevice);
    cudaMemcpy(dev_b, &b, sizeof(int), cudaMemcpyHostToDevice);

    //调用那个要在显卡上面执行的函数。还记得main前面的那个函数么？
    vector_add<<<1, 1>>>(dev_a, dev_b, dev_c);

    //这里把dev_c上面的值复制到内存的变量c上。
    cudaMemcpy(&c, dev_c, sizeof(int), cudaMemcpyDeviceToHost);

    //输出结果
    printf("%d + %d = %d, Is that right?\n", a, b, c);

    //释放显存
    cudaFree(dev_a);
    cudaFree(dev_b);
    cudaFree(dev_c);

    return 0;
}

这就是一个简单的Cuda程序流程，在主机上准备好数据，然后把数据送到显卡，显卡处理好数据后，再把显卡送回到主机内存上。

这个例子非常简单，相信如果你写过几个C语言的程序，很容易就看懂这段代码，毕竟有个几个函数，除了过了个cuda前缀之外，和平常使用的没有很多差别。

你肯定想问，那个__global__是个什么鬼？那个<<<1, 1>>>又是个什么鬼？这些等到下节我们再来解答。