Manycore Programming: CUDA 1

最新推荐文章于 2023-11-07 18:39:47 发布

翻译最新推荐文章于 2023-11-07 18:39:47 发布 · 1k 阅读

文章标签：

#CUDA #manycore #18-645 #cmu #fastcode

CMU: How to Write Fast Code 专栏收录该内容

23 篇文章

订阅专栏

部署运行你感兴趣的模型镜像

多核和多处理器平台的区别
- 作用的区别
- 架构的区别
软硬件的内在模型
应用的语义
- CUDA的存储操作
同步
CUDA平台

在之前介绍了OpenMP的多核编程，这一节主要讲得就是CUDA的多处理器编程。

多核和多处理器平台的区别

作用的区别

这里写图片描述

多核平台的指令以及控制更加复杂，可以在单个周期实现很简单的操作。优化的时候是针对单个的线程，是对线性代码模型的加速。

多处理器更多的资源是在计算部分，实现的是减轻单个处理器的负载。

架构的区别

处理单个数据的能力CPU更强，在最大线程数量、带宽、每个cycle可以处理浮点计算个数来看，GPU更强。

这里写图片描述

什么时候用GPU比较合适呢？当然就是在并行很多的时候采用比较合适。在共享内存很多的时候使用的话，效率会比较低。

软硬件的内在模型

在这里主要是对CUDA做了一些介绍，然后这里的话是谈到了如果想要发挥CUDA的作用，那么必须了解其硬件特性，才能更好发挥其作用。

nVidia的硬件架构：
这里写图片描述

这里面谈到了CUDA的编程模型。语法和C很像。不过定义了主机和从机两部分的代码。主机一般就是CPU、从机就是GPU。主机发送数据到了从机，从机根据kernel编译的指令执行得到数据后返回给主机。

在CUDA的编程模型里面，首先是grid。grid就是比如一个main函数，里面调用了好几个不同的kernel，然后在执行的时候，每个就会生成一个grid。然grid是针对软件而言的，硬件上称之为streaming multiprocessor（SM），然后每个SM内部又有thread block，然后thread block里面又有好几个threads。每个SM之间是相互独立的。每个thread block能够容纳的最大线程数量是一定的。比如现在的gtx1080就是1024个线程。

在软件层面，有一个叫做warp的概念。就是把一些线程集合在一起，方便软件层面的使用。

这里写图片描述

应用的语义

接下来具体介绍CUDA的语法：

__global__ void vcos( int n, float* x, float* y ) { 
    int ix = blockIdx.x*blockDim.x + threadIdx.x; 
    y[ix] = cos( x[ix] );
}

int main() { 
    float *host_x, *host_y; float *dev_x, *dev_y; int n = 1024;
    // CPU的DRAM分配
    host_x = (float*)malloc( n*sizeof(float) ); 
    host_y = (float*)malloc( n*sizeof(float) );
    // GPU的DRAM上分配  
    cudaMalloc( &dev_x, n*sizeof(float) ); 
    cudaMalloc( &dev_y, n*sizeof(float) );

    /* TODO: fill host_x[i] with data here */
    // 数据拷贝 
    cudaMemcpy( dev_x, host_x, n*sizeof(float), cudaMemcpyHostToDevice );

    /* launch 1 thread per vector-element, 256 threads per block */ 
    bk = (int)( n / 256 ); 
    vcos<<<bk,256>>>( n, dev_x, dev_y );

    cudaMemcpy( host_y, dev_y, n*sizeof(float), cudaMemcpyDeviceToHost ); 
    /* host_y now contains cos(x) data */

    return( 0 );
}

CUDA的存储操作

cudaMalloc(void ** pointer, size_t nbytes) 
cudaMemset(void * pointer, int value, size_t count) 
cudaFree(void* pointer)

存储拷贝：

cudaMemcpy( !void *dst, void *src, size_t nbytes,  enum cudaMemcpyKind direction);

direction specifies locations (host or device) of src and dst
Blocks CPU thread:  returns after the copy is completed 
Doesn’t start copying until previous CUDA calls complete enum cudaMemcpyKind 

cudaMemcpyHostToDevice 
cudaMemcpyDeviceToHost 
cudaMemcpyDeviceToDevice
cudaMemcpyPeer

CUDA的函数类型：有global、device、host三种。

这里写图片描述