🎉感谢来自NVIDIA企业级开发者社区的何琨(Ken)老师提供的资料和细致耐心的讲解
异构计算
术语:
Host: CPU和内存(host memory)
Device:GPU和显存(device memory)
CUDA安装
适用设备
所有包含NVIDIA GPU的服务器,工作站,个人电脑,嵌入式设备等电子设备。
软件安装
Windows:https://docs.nvidia.com/cuda/cuda-installation-guide-microsoft-windows/index.html
Windows安装最简单,只需运行一个exe即可。
Linux:https://docs.nvidia.com/cuda/cuda-installation-guide-linux/index.html
具体步骤也可以参考我的这篇博文,但由于时间推移注意替换软件版本号。
Ubuntu20.04下安装Cuda11.0+Nvidia-440+Cudnn7.1.4
Jetson:https://developer.nvidia.com/embedded/jetpack
采用NVIDIA SDK Manager安装。
GPU状态查询:
PC、工作站:nvidia-smi
Jetson:jtop
GPU信息查询:
cd /usr/local/cuda/samples/1_Utilities/deviceQuery
./deviceQuery
CUDA程序的编写模型
把输入数据从CPU内存复制到GPU显存
在执行芯片上缓存数据,加载GPU程序并执行
将计算结果从GPU显存中复制到CPU内存中
关键字
CUDA程序案例
核函数
__global__ void gpu_kernel(int input_M[N * N], int output_M[N * N])
{
int idx = blockIdx.x * blockDim.x + threadIdx.x;
if (idx >= N * N)
return;
if (input_M[idx] > 100)
output_M[idx] = 1;
else if (input_M[idx] < 100)
output_M[idx] = 0;
}
核函数配置及调用
int main(int argc char **argv){
...
//kernel configuration
dim3 dimGrid(ceil(N/TILE_DIM), ceil(M/TILE_DIM), 1);//网格大小配置
dim3 dimBlock(TILE_DIM, BLOCK_ROWS, 1);//块大小配置
//<<<执行配置>>>指定了线程数量以及线程分组
gpu_kernel<<>>(inputMat, outputMat);
...
}
CUDA程序的编译
NVCC使用案例
nvcc --device-c hello_from_gpu.cu -o hello_from_gpu.o
nvcc hello_from_gpu.o hello_cuda_main.cu -o hello_from_gpu
CUDA程序性能分析
NVPROF
Kernel Timeline输出的是以gpu kernel为单位的一段时间的运行时间线,我们可以通过它观察GPU在什么时候有闲置或者利用不够充分的行为,更准确地定位优化问题。nvprof是nvidia提供的用于生成gpu timeline的工具,其为cuda toolkit的自带工具。
可以结合nvvp和nsight进行可视化分析。
NVPROF 使用案例
查看a.exe的运行总体情况:
在不指定参数的情况下默认输出--print-api-summary
执行时可能需要管理员权限
nvprof a.exe
查看a.exe的运行流程:
需要指定参数:--print-api-trace
nvprof --print-api-trace a.exe
实验
编写第一个CUDA程序
- 关键词:\"__global__\" , <<<...>>> , .cu
在当前的目录下创建一个名为hello_cuda.cu的文件,编写第一个Cuda程序:
CPU版本
#include
int main(void)
{
printf("Hello World!\n");
return 0;
}
编译:
nvcc hello_cuda.cu -o hello_cuda
输出:
Hello World!
GPU版本
#include __global__ void cuda_hello(void){ printf("Hello CUDA from GPU!\n");}int main(void){ cuda_hello<<<8,1>>>(); cudaDeviceSynchronize(); return 0;}
编译:
nvcc hello_cuda.cu -o hello_cuda
输出:
Hello CUDA from GPU!Hello CUDA from GPU!Hello CUDA from GPU!Hello CUDA from GPU!Hello CUDA from GPU!Hello CUDA from GPU!Hello CUDA from GPU!Hello CUDA from GPU!
使用nvprof查看程序性能
sudo /usr/local/cuda/bin/nvprof --print-api-summary ./hello_cuda
输出:
==17055== NVPROF is profiling process 17055, command: ./hello_cuda==17055== Warning: Unified Memory Profiling is not supported on the underlying platform. System requirements for unified memory can be found at: http://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#um-requirementsHello CUDA from GPU!Hello CUDA from GPU!Hello CUDA from GPU!Hello CUDA from GPU!Hello CUDA from GPU!Hello CUDA from GPU!Hello CUDA from GPU!Hello CUDA from GPU!==17055== Profiling application: ./hello_cuda==17055== Profiling result: Type Time(%) Time Calls Avg Min Max Name API calls: 99.89% 361.69ms 1 361.69ms 361.69ms 361.69ms cudaLaunchKernel 0.06% 213.18us 97 2.1970us 960ns 58.272us cuDeviceGetAttribute 0.04% 136.67us 1 136.67us 136.67us 136.67us cudaDeviceSynchronize 0.01% 21.152us 1 21.152us 21.152us 21.152us cuDeviceTotalMem 0.00% 10.432us 3 3.4770us 1.5360us 5.4720us cuDeviceGetCount 0.00% 8.0320us 2 4.0160us 2.1440us 5.8880us cuDeviceGet 0.00% 2.9120us 1 2.9120us 2.9120us 2.9120us cuDeviceGetName 0.00% 1.5040us 1 1.5040us 1.5040us 1.5040us cuDeviceGetUuid
尝试利用CUDA API打印GPU参数
device.cu
#include #include #include #include int main(void){ //define variables int deviceCount = 0; int dev = 0, driverVersion = 0, runtimeVersion = 0; //count device cudaError_t error_id = cudaGetDeviceCount(&deviceCount); printf("Detected %d CUDA Capable device(s)\n", deviceCount); //select device cudaSetDevice(dev); cudaDeviceProp deviceProp; cudaGetDeviceProperties(&deviceProp, dev); printf("Device %d: \"%s\"\n", dev, deviceProp.name); //CUDA info cudaDriverGetVersion(&driverVersion); cudaRuntimeGetVersion(&runtimeVersion); printf("CUDA Driver Version / Runtime Version %d.%d / %d.%d\n", driverVersion / 1000, (driverVersion % 100) / 10, runtimeVersion / 1000, (runtimeVersion % 100) / 10); printf("CUDA Capability Major/Minor version number: %d.%d\n", deviceProp.major, deviceProp.minor); return 0;}
编译:
nvcc device.cu -o device
输出:
Detected 1 CUDA Capable device(s)Device 0: "Xavier"CUDA Driver Version / Runtime Version 10.2 / 10.2CUDA Capability Major/Minor version number: 7.2