目标
- 编写第一段CUDA C代码
- 了解主机(Host)端编写的代码和设备端(Device)编写的代码的区别
- 如何从主机上运行设备端代码
- 了解如何在支持CUDA的设备上使用内存
- 了解如何查询系统中支持CUDA的设备信息
第一个CUDA C程序
CUDA C在很大程度上与标准C没有区别。在GPU上执行的函数通常称为核函数。
#include <iostream>
__global__ void kernel(void) {
}
int main()
{
kernel<<<1,1>>>();
printf("Hello,World\n");
return 0;
}
这个程序中,有一个global修饰的核函数kernel(),调用它时,用了修饰字符<<<1,1>>>,global字符表示这个函数在设备端运行而不是在主机端。
代码来源于书籍
#include <iostream>
__global__ void add(int a,int b,int *c){
*c = a+b;
}
int main(void)
{
int c;
int *dev_c;
HANDLE_ERROR(cudaMalloc((void**)&dev_c,sizeof(int)));
add<<<1,1>>>(2,7,dev_c);
HANDLE_ERROR(cudaMemcpy(&c,dev_c,sizeof(int),cudaMemcpyDeviceTohost));
printf("2+7=%d\n",c);
cudaFree(dev_c);
return 0;
}
在设备端进行运算操作时,有要保存的数据,必须要先在设备端分配空间,然后运算完之后再传输回主机端保存。记住要释放设备端空间。设备端cudaMalloc返回的指针不一定能进行引用。
cudaMalloc()函数是分配设备端内存的,这个函数类似于malloc(),第一个参数是一个指针,指向保存变量的地址,第二个参数是大小。HANDLE_ERROE是自定义的宏,判断是否出错。cudaMemcpy()将设备端数据复制到主机端。
总结如下:
1. 可以将cudaMalloc分配的指针传递给设备上执行的函数。
2. 可以在设备代码中使用cudaMalloc分配的指针进行操作。
3. 可以将cudaMalloc分配的指针传递给主机上执行的函数。
4. 不能在主机代码中使用cudaMalloc分配的指针进行内存读写操作。
访问设备内存两种最常见的方法就是在设备代码中使用设备指针和调用cudaMemcpy函数。
cudaMemcpy函数类似于memcpy函数,只是在多了一个参数来指定设备内存究竟是源指针还是目标指针,它可以是cudaMemcpyDeviceToHost,cudaMemcpyHostToDevice和cudaMemcpyDeviceToDevice。
查询设备
利用CUDA C 可以查询到的信息
设备属性 | 描述 |
---|---|
char name[256] | 标识设备的ASCII字符串 |
size_t totalGlobalMem | 设备上全局内存的总量(字节) |
size_t sharedMemPerBlock | 线程块中可以是哟你哦个的最大共享内存数量(字节) |
int regsPerBlock | 每个线程中可以用的32位寄存器数量 |
int warpSize | 在一个线程束中包含的线程数量 |
size_t memPitch | 在内存复制中最大的修正量(字节) |
int amxThreadsPerBlock | 在一个线程块中可以包含最大的线程数量 |
int maxThreadsDim | 多维线程块数组中,每一维可以包含的最大线程数量 |
size_t totalConstMem | 常量内存的总量 |
int major | 设备计算功能集的主版本号 |
int minor | 设备功能集的次版本号 |
size_t textureAlignment | 设备的纹理对齐要求 |
int deviceOverlap | 一个布尔类型的值,表示设备是否可以同时执行一个cudaMemory()调用和一个核函数调用 |
int multiProcessorCount | 设备上多处理器的数量 |
int kernelExecTimeoutEnabled | 一个布尔值,表示在该设备上执行的核函数是否存在运行时限制 |
int integrated | 表示设备是否是一个继承GPU |
int canMapHostMemory | 一个布尔类型,表示设备是否将主机内存映射到CUDA设备地址空间 |
int computeMode | 表示设备的计算模式:默认(default),独占(Exclusive),禁止(Prohibited) |
int maxTexture1D | 一维纹理的最大大小 |
int maxTexture2D[2] | 二维纹理的最大维数 |
int maxTexture3D[3] | 三维纹理的最大维数 |
int maxTexture2DArray[3] | 二维纹理数组的最大维数 |
int concurrentKernels | 表示设备是否支持在同一个上下文中同时执行多个核函数 |
设备查询代码
#include <stdio.h>
int main()
{
cudaDeviceProp prop;
int count;
HANDLE_ERROR(cudaGetDeviceCount(&count));
for(int i = 0;i<count;i++)
{
HANDLE_ERROR(cudaGetDeviceProperties(&prop,i));
//对设备数量执行某些操作
}
return 0;
}
欢迎评论,一起学习。