CUAD笔记
CUDA学习
1.环境配置
CUDA环境的搭建在windows10下需要安装vs2013+cuda7.5 工具包,linux下安装工具包后需要再安装gcc等编译环境并配置环境变量,mac类似linux,只不过编译环境可以直接使用xcode。
2.编译
物理运行方式: 发布模式:nvcc <文件名>.cu [-o 输出文件名] 调试模式(只能调试主机代码,不能调试gpu代码):nvcc –g <文件名>.cu 模拟运行方式: 无符号模式(所有代码运行于主机):nvcc–deviceemu <文件名>.cu 有符号模式(所有代码运行于主机):nvcc–deviceemu -g <文件名>.cu
3.内存管理
首先分配内存,然后从主机内存向gpu内存拷贝,然后进行计算,将结果拷贝回主机内存,最后释放不需要的内存。
1.分配内存与释放内存
函数原型 (1) cudaMalloc(void ** 指针,size_t 字节数) (2) cudaMemset(void * 指针,int值,size_t 设置数目) (3) cudaFree(void * 指针)
int n=1024;
int nbytes =1024*sizeof(int);
int *d_a=0;
cudaMalloc((void **) &d_a, nbytes);
cudaMemset(d_a,0,nbytes);
cudaFree(d_a);
2.数据复制
函数原型
cudaMemcpy(void * dst,void * src,size_t nbytes,enum cudaMemcpyKind direction)
disection是用于标示复制设备的值,无论是多少,此api均将把src指向的空间内的数据复制到dst指向的空间中,调用此api的cpu线程将在复制完成后返回,此api不能在GPU中调用。
(1)enum cudaMemcpyKind解释
cudaMemcpyHostToDevice表示从主机复制到设备,实际上是从内存复制到显存 cudaMemcpyDeviceToHost表示从设备复制到主机,实际上是从显存复制到内存 cudaMemcpyDeviceToDevice表示从设备复制到设备,实际上用于两个显卡间的显存复制或两个显存空间间的复制
int main(void)
{
float *a_h,*b_h;//主机数据
float *a_d,*b_d;//设备数据
int N=14,nBytes,I;
nBytes=N*sizeof(float);
a_h=(float *)malloc(nBytes);
b_h=(float *)malloc(nBytes);
cudaMalloc((void**)&a_d,nBytes);
cudaMalloc((void**)&b_d,nBytes);
for(i=0,i<N;i++)
{
a_h[i]=100.0f+I;
}
cudaMemcpy(a_d,a_h,nBytes,cudaMemcpyHostToDevice);
cudaMemcpy(b_d,a_d,nBytes,cudaMemcpyDeviceToDevice);
cudaMemcpy(b_h,b_d,nBytes,cudaMemcpyDeviceToHost);
for(i=0; i<N;i++)
{
assert(a_h[i]==b_h[i]);
}
free(a_h);
free(b_h);
cudaFree(a_d);
cudaFree(b_d);
return 0;
}
4.核函数(内核函数)
1.内核函数是一种有一些限制的函数:
1)不能访问主机内存
2)返回类型只能是void
3)不允许可变参数
4)不允许递归
5)不允许静态变量
2.函数参数将自动地从主机复制到设备
- __ global__
(1)此种函数从主机调用,在设备上执行
(2)只允许void返回 New Roman
2)device
(1)此种函数从设备上调用,在设备上执行
(2)不允许被主机代码调用
3)host
此种函数从主机调用,在主机上执行(如不写修饰语,默认函数为__host__修饰)
4)说明
(1)dG表明grid(网格)中blocks(块)的数目
<1>只有两个维度有效,x和y,z始终只能为1(当前限制,不知道以后会不会解除)
<2>块的总量:dG.x*dG.y
(2)dB表明块中threads(线程)的数目
<1>三个维度都有效,x、y和z
<2>线程的总量:dB.x*dB.y*dB.z
(3)dim3中未被定义的值默认为1
dim3 grid,block;
grid.x=2;
grid.y=4;
block.x=8;
block.y=16;
kernel<<<grid,block>>>
dim3 grid(2,4),block(8,16);
kernel<<<grid,block>>>
kernel<<<32,512>>>
- 设备变量管理
(1)所有设备上运行的函数都能*访问它们的局部变量
(2)特殊变量
<1>dim3 gridDim:网格中块的数目(最多二维)
<2>dim3blockDim:块中线程数目
<3>dim3blockIdx:在网格中的块引索(运行时自动生成的唯一值)
<4>dim3threadIdx:块中的线程引索(运行时自动生成的唯一值)
-
主机同步
(1)所有核函数的运行都是异步的<1>主机调用后会立即返回
<2>内核函数将在cuda之前的calls完成后执行
(2)cudaMemcpy()是个同步过程
<1>将会在复制完成后返回主机
<2>复制在cuda之前的calls完成后执行
<3>使用方式:先复制数据到设备,调用核函数,执行其他主机代码,从设备复制结果数据
(3)cudaThreadSynchronize()
执行到此处后等待块中所有线程均运行到此处后继续执行
7)变量修饰语
(1)device
<1>在设备全局内存中,容量大,所有线程均可访问,无缓存
<2>使用cudaMalloc进行分配(隐含__device__修饰)
<3>存在于整个程序生存周期内(你不释放或程序结束都在那)
(2)shared
<1>块内线程可访问,容量小,速度快
<2>存在于块线程存在时(块线程结束自动释放)
(3)未修饰变量
<1>标量和运行时建立的矢量在寄存器中
<2>无法安置在寄存器中的变量保存在local(本地)内存中
8)设备同步
(1)void __syncthreads()
<1>设备上调用
<2>只有块中所有线程均执行到此处时才可继续向下执行
(2)分支操作
只有处于一个分支的所有线程执行完毕才会执行另一个分支的线程
9)cuda错误反馈
(1)除调用核函数外的所有cuda函数调用都会返回错误代码,类型为cudaError_t,未发生错误返回cudaSuccess。
(2)cudaError_t cudaGetLastError(void)
<1>返回最后一个错误的错误代码(如果没有错误也会返回一个值,值为cudaSuccess)
<2>可以用于获取核函数调用时发生的错误的错误代码
(3)char* cudaGetErrorString(cudaError_t)
用于将错误代码转化为文本信息 p