CUDA 编程——Vector Addition

1 准备工作

1）安装 xshell

2）远程连接服务器，打开 jupyter 服务

3）网页进入对应的 jupyter 服务

2 Launcher 介绍

3 查看 GPU 设备信息

1）点击 Text File

2）在其中输入对应的代码
因为 Tesla K 系列都是采用 Kepler 架构，所以每一个 SM 中的 cuda core 的数量为 192。

#include <stdio.h>
#include <time.h>
#include <math.h>
#include "cuda_runtime.h"
#include "device_launch_parameters.h"


int main()
{
    cudaDeviceProp deviceProp;
    cudaGetDeviceProperties(&deviceProp, 0);
    printf("设备名称与型号: %s\n", deviceProp.name);
    printf("显存大小: %d MB\n", (int)(deviceProp.totalGlobalMem / 1024 / 1024));
    printf("含有的SM数量: %d\n", deviceProp.multiProcessorCount);
    printf("CUDA CORE数量: %d\n", deviceProp.multiProcessorCount * 192);
    printf("计算能力: %d.%d\n", deviceProp.major, deviceProp.minor);
}

3）重命名文件，后缀名改为 .cu

4）在 Terminal 中输入以下指令编译代码

nvcc filename.cu -o filename.out

5）在 Terminal 中输入一下指令运行可执行文件

./filename.out

4 向量加法的实现

1）点击 Text File

2）在其中输入对应的代码
CPU version

#include <stdio.h>
#include <time.h>

const int N = 50000;
int main()
{
    clock_t start, end;
    start = clock();
    float a[N], b[N], c[N];
    for (int i = 0; i < N; i ++) a[i] = 0, b[i] = 0;  // initial
    for (int i = 0; i < N; i ++) c[i] = a[i] + b[i];  // vector addition
    end = clock();
    printf("Vector addition on CPU use %.8f s.\n", (float)(end - start) / CLOCKS_PER_SEC);
}

CUDA version

#include <stdio.h>
#include <time.h>
#include <math.h>
#include "cuda_runtime.h"
#include "device_launch_parameters.h"

const int N = 50000;
__global__ void additionKernelVersion1(float*, float*, float*, const int);
__global__ void additionKernelVersion2(float*, float*, float*, const int);
__global__ void additionKernelVersion3(float*, float*, float*, const int);
int main()
{
    clock_t start, end;
    start = clock();
    float a[N], b[N], c[N];
    for (int i = 0; i < N; i ++) a[i] = 0, b[i] = 0;  // 在 host 端初始化数据
    float *device_a, *device_b, *device_c = NULL;  
    cudaMalloc((void**)&device_a, sizeof(float) * size);  // 在 device 分配内存
    cudaMalloc((void**)&device_b, sizeof(float) * size);  // 在 device 分配内存
    cudaMalloc((void**)&device_c, sizeof(float) * size);  // 在 device 分配内存
    cudaMemcpy(device_a, a, sizeof(float) * size, cudaMemcpyHostToDevice);  // 将 host 的数据拷贝到 device
    cudaMemcpy(device_b, b, sizeof(float) * size, cudaMemcpyHostToDevice);  // 将 host 的数据拷贝到 device
    additionKernelVersion1<<<ceil(N / 32), 32>>>(device_a, device_b, device_c, size);  // 使用 kernel 进行运算
    // additionKernelVersion2<<<ceil(N / 32), 32>>>(device_a, device_b, device_c, size);  // 使用 kernel 进行运算
    // additionKernelVersion3<<<ceil(N / 32), 32>>>(device_a, device_b, device_c, size);  // 使用 kernel 进行运算
    cudaMemcpy(device_c, c, sizeof(float) * size, cudaMemcpyDeviceToHost);  // 将 device 中的计算结果拷贝到 host
    cudaFree(device_a);  // 释放 device 中的内存
    cudaFree(device_b);  // 释放 device 中的内存
    cudaFree(device_c);  // 释放 device 中的内存
    end = clock();
    printf("Vector addition version 1 on GPU use %.8f s.\n", (float)(end - start) / CLOCKS_PER_SEC);
    // printf("Vector addition version 2 on GPU use %.8f s.\n", (float)(end - start) / CLOCKS_PER_SEC);
    // printf("Vector addition version 3 on GPU use %.8f s.\n", (float)(end - start) / CLOCKS_PER_SEC);
}
__global__ void additionKernelVersion1(float* A, float* B, float* C, const int size)
{
    int i = blockIdx.x * blockDim.x + threadIdx.x;
    C[i] = A[i] + B[i];
}
__global__ void additionKernelVersion2(float* A, float* B, float* C, const int size)
{
    int i = (blockIdx.x * blockDim.x + threadIdx.x) * 2;
    C[i] = A[i] + B[i];
    C[i + 1] = A[i + 1] + B[i + 1];
}
__global__ void additionKernelVersion3(float* A, float* B, float* C, const int size)
{
    int i = blockIdx.x * blockDim.x + threadIdx.x;
    C[i] = A[i] + B[i];
    C[i + blockDim.x] = A[i + blockDim.x] + B[i + blockDim.x];
}

3）重命名文件，后缀名改为 .cu

4）在 Terminal 中输入以下指令编译代码

nvcc vector_addition.cu -o vector_addition.out

5）在 Terminal 中输入一下指令运行可执行文件

./vector_addition.out