cuda_c学习笔记-向量加法

用cuda计算向量加法A+B=C

流程：

1.申请主机内存。向量A，向量B，计算结果C

2.初始化数据。用0-1之间的随机数初始化向量A，B，C

3.GPU内存申请。申请A，B，C需要的GPU内存空间

4.数据拷贝。把数据从主机内存拷贝至GPU内存

5.计算需要的线程数和线程块数。

6.调用GPU加法函数

7.数据拷贝。把结果从GPU内存拷贝至主机内存。

8.在CPU上重新运行一遍，与GPU结果进行对照。

9.释放GPU内存。

10.释放主机内存。

11.重置GPU状态。

要点：内存管理，数据拷贝。

代码：

#include <stdio.h>
#include <cuda_runtime.h>
__global__ void
vectorAdd(const float *A, const float *B, float *C, int numElements)
{
    int i = blockDim.x * blockIdx.x + threadIdx.x;

    if (i < numElements)
    {
        C[i] = A[i] + B[i];
    }
}

int main(void)
{
    //检测cuda返回值
    cudaError_t err = cudaSuccess;

    //初始化向量维度
    int numElements = 50000;
    //计算内存需求
    size_t size = numElements * sizeof(float);

    printf("[Vector addition of %d elements ]
", numElements);

    //
    // 对主机的A,B,C申请内存空间
    float *host_A = (float *)malloc(size);
    float *host_B = (float *)malloc(size);
    float *host_C = (float *)malloc(size);
    //判断是否申请成功
    if (host_A == NULL || host_B == NULL || host_C == NULL)
    {
        fprintf(stderr, "Failed to allocate host vectors!
");
        exit(EXIT_FAILURE);
    }
    // 初始化主机A,B
    for (int i = 0; i < numElements; ++i)
    {
        host_A[i] = rand()/(float)RAND_MAX;
        host_B[i] = rand()/(float)RAND_MAX;
    }

    //
    //申请cuda内存空间并判断
    float *device_A = NULL;
    err = cudaMalloc((void **)&device_A, size);
    if (err != cudaSuccess)
    {
        fprintf(stderr, "对向量A申请cuda内存空间失败 (错误代码 %s)!
", cudaGetErrorString(err));
        exit(EXIT_FAILURE);
    }
    float *device_B = NULL;
    err = cudaMalloc((void **)&device_B, size);
    if (err != cudaSuccess)
    {
        fprintf(stderr, "对向量B申请cuda内存空间失败 (错误代码 %s)!
", cudaGetErrorString(err));
        exit(EXIT_FAILURE);
    }
    float *device_C = NULL;
    err = cudaMalloc((void **)&device_C, size);
    if (err != cudaSuccess)
    {
        fprintf(stderr, "对向量C申请cuda内存空间失败 (错误代码 %s)!
", cudaGetErrorString(err));
        exit(EXIT_FAILURE);
    }
    //拷贝数据到cuda内存并检测
    printf("Copy input data from the host memory to the CUDA device
");
    err = cudaMemcpy(device_A, host_A, size, cudaMemcpyHostToDevice);
    if (err != cudaSuccess)
    {
        fprintf(stderr, "Failed to copy vector A from host to device (error code %s)!
", cudaGetErrorString(err));
        exit(EXIT_FAILURE);
    }
    err = cudaMemcpy(device_B, host_B, size, cudaMemcpyHostToDevice);
    if (err != cudaSuccess)
    {
        fprintf(stderr, "Failed to copy vector A from host to device (error code %s)!
", cudaGetErrorString(err));
        exit(EXIT_FAILURE);
    }

    //计算线程块与线程
    //每线程块线程数
    int threadsPerBlock = 256;
    //每网格线程块数
    int blocksPerGrid = (numElements + threadsPerBlock - 1)/ threadsPerBlock;
    printf("CUDA kernel launch with %d blocks of %d threads
", blocksPerGrid, threadsPerBlock);
    vectorAdd<<<blocksPerGrid, threadsPerBlock>>>(device_A, device_B, device_C, numElements);
    //判断cuda程序运行情况
    err = cudaGetLastError();
    if (err != cudaSuccess)
    {
        fprintf(stderr, "Failed to launch vectorAdd kernel (error code %s)!
", cudaGetErrorString(err));
        exit(EXIT_FAILURE);
    }

    //计算结果拷贝回主机
    printf("Copy output data from the CUDA device to the host memory
");
    err = cudaMemcpy(host_C, device_C, size, cudaMemcpyDeviceToHost);
    if (err != cudaSuccess)
    {
        fprintf(stderr,"计算结果拷贝回主机失败（错误代码：%s）
",cudaGetErrorString(err));
        exit(EXIT_FAILURE);
    }

    //结果验证
    for(int i = 0; i < numElements; ++i)
    {
        if(fabs(host_A[i] + host_B[i] - host_C[i]) > 1e-5)
        {
            fprintf(stderr,"验证失败%d
",i);
            exit(EXIT_FAILURE);
        }
    }
    printf("验证成功
");

    //释放cuda内存和主机内存
    err = cudaFree(device_A);
    err = cudaFree(device_B);
    err = cudaFree(device_C);
    free(host_A);
    free(host_B);
    free(host_C);

    //重置cuda状态
    err = cudaDeviceReset();
    printf("结束");
    return 0;
}

好烦……GPU搞起来真麻烦