前言
由于最近忙着秋招,本系列博客最近只记录下自己学CUDA的例子: 这里简单说下代码的计算流程:
具体步骤如下:
主机内存分配:在 CPU 上分配三个数组 ha、hb 和 hc。
初始化数组:将 ha 和 hb 的所有元素分别初始化为常量 a 和 b。
设备内存分配:在 GPU 上分配三个数组 da、db 和 dc。
数据传输:将主机上的数组 ha 和 hb 复制到设备上的数组 da 和 db。
定义网格和块尺寸:设置 CUDA 内核的执行配置。
启动内核:在 GPU 上并行执行向量加法。
数据传输:将设备上的结果数组 dc 复制回主机上的数组 hc。
释放内存:释放主机和设备上分配的所有内存。
1、cuda源码
#include <stdio.h>
const double a = 1.23;
const double b = 2.34;
void __global__ add(const double *x, const double *y, const double *z,const int N );
void __global__ add(const double *x, const double *y, double *z, const int N)
{
const int tid = blockDim.x * blockIdx.x + threadIdx.x;
if(tid < N)
{
z[tid] = x[tid] + y[tid];
}
}
int main()
{
const int N = 1000;
const int M = sizeof(double) * N;
double *ha = (double *)malloc(M);
double *hb = (double *)malloc(M);
double *hc = (double *)malloc(M);
// assignment
for(int i=0; i < N; ++i)
{
ha[i] = a;
hb[i] = b;
}
//
double *da, *db, *dc;
cudaMalloc((void**)&da, M);
cudaMalloc((void**)&db, M);
cudaMalloc((void**)&dc, M);
cudaMemcpy(da,ha,M,cudaMemcpyHostToDevice);
cudaMemcpy(db,hb,M,cudaMemcpyHostToDevice);
// kernel fun
const int block_size = 128;
const int grid_size = (N + block_size -1)/ block_size;
add<<<grid_size,block_size>>>(da,db,dc,N);
cudaMemcpy(hc,dc,M,cudaMemcpyDeviceToHost);
free(ha);
free(hb);
free(hc);
cudaFree(da);
cudaFree(db);
cudaFree(dc);
return 0;
}
编译指令
nvcc -arch=sm_75 add.cu -o add
2、变量解释
在cuda中,应用最多的可能就是grid_size和block_size,以上面例子为例:假设Tesla V100的block_size = 100,当需要并行计算1000个数字加法时,那么我们就需要用到:
g
r
i
d
_
s
i
z
e
=
N
+
b
l
o
c
k
_
s
i
z
e
−
1
b
l
o
c
k
_
s
i
z
e
grid\_size = \frac{N+block\_size-1}{block\_size}
grid_size=block_sizeN+block_size−1
也就是需要申请最少覆盖完1000个数字的block个数。
总结
这段代码的主要目的是在一个 GPU 上并行地对两个向量进行逐元素加法,并将结果存储在第三个向量中。另外,简单解释了cuda中常用的block_size和grid_size的申请计算方法。
348





