7月4日,2022 CUDA on Arm Platform线上训练营开始第一天的课程。
第一天的课程,NVIDIA开发者社区何琨老师重点讲解:
- 基于Arm的Jetson开发环境介绍,Arm Linux系统简介(1.1理论课 实验课)
介绍实验平台,介绍Linux编译的基本技巧,介绍基本的开发环境。实验课:Makefile 编写规范。
- GPU架构及异构计算(1.2)
- 介绍GPU架构以及异构计算的基本原理
- 介绍GPU硬件平台
- 介绍基于Arm的嵌入式平台GPU架构和编程模型之间的关系,介绍Grace CPU相关
- 最新的GPU应用领域,GPU在现代计算机科学中的通用性
- CUDA编程模型---初识CUDA(1.3 1.4实验课)
- CUDA程序的编译
- GPU线程的调用
- GPU和CPU的通讯
- 使用多个线程的核函数
- 使用线程索引
- 多维网络
- 网格与线程块
- 利用NVProf查看程序执行情况
- 实验课内容:编写程序HelloCUDA,并且利用NVCC编译。编写VectorAdd多线程程序,和多维网络线程程序,并用nvprof来测试不同设置情况下运行速
课堂问题汇总:
1. __device__ 的返回类型可以不是void?
--可以的。但这种情况下,执行的效果可能需要其他方式返回了。(例如通过一个参数int *p, 然后写入p指向的空间),
2.一般返回值都通过指针参数返回比较好?
--都行,常见的习惯是标量值,可以直接作为返回值,每个线程超过1个值(例如需要返回10个float),则建议用指针。
3. 和CPU上只能同时执行有限数量(常见的例如8核16个超线程的CPU)的十几个、几十个。GPU上能同时执行海量的线程数量,例如几十万、上百万。可以有效的发挥GPU设备的能力。而如何有效的管理这么大数量的线程,则需要"线程组织形式", 可以有效的管理、执行问题,避免混乱。
4. 来自Fortran同学注意了,我们的GPU上的CUDA C语言,采用的是下标从0开始,而不是1.
5. 学校:1个grid /年级:1个block/班:1个block/你:1个线程
6. 指定超过SM中core数量的话会怎样?无问题的。你可以理解成每个SM都是海量超线程的。例如我们本次例子的Jetson设备,1个SM只有128个SP,却可以同时执行2048个线程(你可以理解成16倍超线程)。再多也是可以的,用其他方式继续调度
7. 线程数目可以远大于物理core数目
8. 1个block在一个sm里面执行,sm是什么?
--一般情况下,可以直接将GPU的SM理解成CPU的一个物理核心. 按SM划分有好多好处。例如一个GPU可以简单的通过横向扩充SM,即可扩大规模。例如1个block的线程限定给1个SM,可以让1个block的线程在SM内部高效的执行数据交换/交流/同步之类的。
9. 写cuda程序的时候能申请的最大线程数不是无限的, 最大的线程数量:1024*(2^31-1)*65535*65535
10. 一个block有多少个线程是调用的时候自己指定的?而不是固定的?是自己(你)定的。
11. 如果两个进程运行,调用的函数都同时使用同一个blockid和threadid,会不会有冲突的?
--不会。依然各自是各自的线程(虽然两次启动线程的编号有重复的)。
12. 不能直接将一次kernel启动理解成1个CPU上的process的。两回事。你理解成“一次能开辟很多线程的函数调用较好”。
13. 如果cuda申请的thread不足了,调用的函数会怎么样??就是报错如何处理? --如果你指定了超多的启动规模,超出了你卡的能力,会报告“无效启动配置”。
14. cudaDeviceSynchronize();是同步哪个步骤呢,是不同block的计算结果么?
--CPU上的调用者等待GPU上的之前的所有进行中的异步任务完成。和GPU上的blocks之间互相同步(那个叫全局同步)无关
以下是学员学习的笔记分享: