【参加CUDA线上训练营】-Day1*

【参加CUDA线上训练营】-Day1

What is CUDA?

CUDA是NVIDIA开发的一种并行计算平台和编程模型，用于在其自己的GPU（图形处理单元）上进行大规模简易指令运算即SIMD。 从而使开发人员能够利用GPU的能力来实现计算的可并行化部分和与CPU工作协同，最终加快大规模数组矩阵运算的速度。

What‘s the Interface of CUDA

CUDA并不是独立的软件或硬件，他是NVIDIA构筑的软硬件捆绑生态体系中最重要的一环。他实体体现在自家GPU显卡，在软件层面提供了多种开发语言内联调用（C/C++/Python/Java）等等。CUDA向下对应着自家GPU平台实现任务分发和并行计算。向上提供者操作系统与高级语言接口，通过构筑闭环的生态和开放开发环境。使得CUDA成为科学计算与游戏渲染的通用框架，保持了产品在未来的竞争力。

What’s the developing platform

本次训练营采用了云教学的方式，有jetson nano和xarvier nx各50个。这俩平台熟悉的朋友应该知道是边缘计算与ROS系统控制的常用设备，他们均属于Tegra系列机。Tegra中文名图睿是基于ARM架构通用处理器品牌（“Computer on a chip”片上计算机），能够为便携设备提供高性能、低功耗体验。使用jupyter notebook 和虚拟ssh远程连接，实现教学活动。

What’s the differences between Jetson NANO and traditional GPUS

1. NANO集成了CPU、内存、GPU、外置TF卡和IO通道等等。功能完备，就是一台小型机。可独立完成计算推理任务。
2. 传统大型GPU仅具备图像处理功能，是CUDA编程模型中的Device设备，需要使用PCIE通道与HOST（cpu、DRAM）进行通信才可工作不能独立运行
3. NANO最低功耗仅10W即可运行，单显卡功率已破百瓦，在户外边缘场景NANO更合适
4. NANO性能也相对较低，但相较于传统CPU其并行能力是百倍，有力提高了边缘图像处理能力

CUDA Programming Model

*
通过学习编程模型，CUDA主要由HOST与DEVICE设备编程组成，通过 cudaMemcpy（）函数传送数据，cudaMalloc（）分配显存，<<<>>>修饰符为并行化函数建立Block和线程。

通过block与线程实现并行计算，当GPU中的SM采用的对应Block与Thread架构，warp(线程束)是最基本的执行单元，一个warp包含32个并行thread，一个SM的CUDA core会分成几个warp（即CUDA core在SM中分组)，由warp scheduler负责调度。所以我们实际上最多可同时执行32分thread，这也是为什么block中的线程数最好为32的倍数，这样可以充分的利用好运算核心，如果不是，那最后那一部分使用不足32个线程，会使得gpu利用率降低。

How to separate jobs on CPU and GPUS

1. __global__执行空间说明符将函数声明为内核。它的功能是：
   在设备上执行可从主机调用，可在计算能力为3.2或更高的设备调用
   __global__函数必须具有void返回类型,同时异步运算。
2. device:device执行空间说明符声明了一个函数：在设备上执行，只能从设备调用。__global__和__device__执行空间说明符不能一起使用。
3. host:__host__执行空间说明符声明了一个函数：在主机上执行，只能从主机调用。

Makefile organize a huge c project

本质：将Shell组织成脚本，方便工程管理和后期修改
为了方便文件组织和梳理调用关系，使用Makefile描述文件结构与层级关系。

CUDA Programming two Vectors array plus examples

#include <stdio.h>
#include <math.h>

void __global__ add(double *x,double *y,double *z,int count)
{
    int n = blockDim.x * blockIdx.x + threadIdx.x;
    if(n<count)
    {
        z[n]=x[n]+y[n];
    }

}

void check(double *z,int N)
{
    bool error = false;
    for(int n= 0;n < N;++n)
    {
        if(fabs(z[n]-3)>(1.0e-10))
            error = true;
    }
    printf("%d",error);

}

int main()
{
    int N = 1000;
    int M = sizeof(double)*N;
    double *h_x = (double*) malloc(M);
    double *h_y = (double*) malloc(M);
    double *h_z = (double*) malloc(M);
    
    for (int n = 0; n < N; ++n)
    {
        h_x[n] = 1;
        h_y[n] = 2;
    }
    
    double *d_x, *d_y, *d_z;
    cudaMalloc((void**)&d_x,M);
    cudaMalloc((void **)&d_y, M);
    cudaMalloc((void **)&d_z, M);
    cudaMemcpy(d_x, h_x, M, cudaMemcpyHostToDevice);
    cudaMemcpy(d_y, h_y, M, cudaMemcpyHostToDevice);
    
    
    const int block_size = 128;
    const int grid_size = (N + block_size - 1) / block_size;
    add<<<grid_size, block_size>>>(d_x, d_y, d_z, N);
    
    cudaMemcpy(h_z, d_z, M, cudaMemcpyDeviceToHost);
    check(h_z, N);

    free(h_x);
    free(h_y);
    free(h_z);
    cudaFree(d_x);
    cudaFree(d_y);
    cudaFree(d_z);
    return 0;
    
}

在代码中我们使用h_x,d_x，表明数据为谁创建。使用cudaMalloc（）分配显存，cudaMemcpy（）在Device和Host之间拷贝数据。传输完成使用GPU多线程 add<<<grid_size, block_size>>>(d_x, d_y, d_z, N);实现并行加法，最后接收GPU计算数据并检查校验数据check（）是否有误。

Learning Feedback

课程容量很大，老师对专业内容非常熟悉，不乏优秀学员。CUDA编程对计算机组成原理，C语言，数据结构，线性代数等专业基础课相关性较高。需要学习者不断利用已学知识来切身体会实践。