【C++】CUDA核函数

1. 从C++到CUDA编程

cuda中的Hello World步伐
nvcc：

• 安装CUDA即可使用nvcc
  
• nvcc支持纯C++代码的编译
  
• 编译扩展名为.cu的CUDA文件
  
• 编译CUDA文件命令：nvcc hello.cu -o hello
  

使用nvcc编译步伐时，会将纯粹的C++代码交给C++编译器，自己只处理属于CUDA的代码

1. // hello.cu (此代码并没有调用GPU设备)
2. #include <stdio.h>
3. int main(){
4.     printf("Hello World\n");
5.     return 0;
6. }

复制代码

留意：在cuda编程中，不能使用cout进行输出，只能使用printf
2. 核函数

核函数简介：

• 核函数是运行在GPU上的函数，可以并行实验，核函数通过特别的语法和关键字定义，而且是 GPU 编程的核心。
  

定义核函数：

• 必须使用限定词__global__修饰
  
• 核函数返回值必须是void
  
• 核函数通常以<<<blockNum, threadNum>>>形式来设置实验线程和块
  

核函数留意事项：

• 核函数只能访问GPU内存
  
• 核函数不能使用变长参数
  
• 核函数不能使用静态变量
  
• 核函数不能使用函数指针
  
• 核函数具有异步性
  

核函数编写流程：

1. int main(){
2.         主机代码
3.         核函数调用
4.         主机代码
5.         return 0;
6. }

复制代码

代码示例：

1. #include <stdio.h>
3. __global__ void hello_from_gpu(){
4.     printf("Hello World from GPU \n");
5. }
7. int main(){
8.     hello_from_gpu<<<1,6>>>();
9.     // 同步主机与设备，促使缓冲区刷新，打印信息到终端。该函数会等待GPU设备执行完毕
10.     cudaDeviceSynchronize();
11.     return 0;
12. }

复制代码

代码解析：
1.核函数hello_from_gpu

• 核函数通过__global__限定符定义
  
• 函数内使用printf 打印 Hello World from GPU
  

2.核函数启动

• 在main函数中，hello_from_gpu<<<1,6>>>();启动了核函数
  
• 这里的<<<1,6>>>表现启动1个线程块，每个块包罗6个线程
  

3.同步

• 使用cudaDeviceSynchronize();来同步主机与设备，确保 GPU 上的全部线程实验完毕后，主机步伐才继承实验
  

代码输出：

1. Hello World from GPU
2. Hello World from GPU
3. Hello World from GPU
4. Hello World from GPU
5. Hello World from GPU
6. Hello World from GPU

复制代码

留意：
   由于核函数的实验是异步的，cudaDeviceSynchronize() 负责期待 GPU 实验完毕并确保全部输出都能被打印出来。如果没有这个同步步骤，可能会导致输出内容丢失，大概步伐提前竣事。
  总结：

• 通过 __global__修饰符定义的核函数在 GPU 上并行实验。通过 <<<blocks, threads>>> 语法设置并行盘算的块和线程数。
  
• 核函数实验是异步的，主机步伐不会期待设备代码实验完毕，通常使用 cudaDeviceSynchronize() 来期待 GPU 完成操作。
  
• 在 CUDA 中，printf 用于设备端输出，主机端的 cout 无法使用。
  

3. 题目与思索

3.1 CUDA编程中.cu文件作用是什么

   在CUDA编程中，.cu文件是CUDA源码文件的尺度扩展名。CUDA代码分为两部门：一部门运行在主机CPU上，另一部门运行在GPU上。.cu文件用于包罗这些代码，而且他们的内容会被nvcc编译器处理
  

• 主机代码（C++代码）使用尺度C++编译器进行编译
  
• 设备代码（CUDA核函数）由nvcc进行编译，生成与CUDA设备兼容的代码
  

  3.2 核函数为什么不能使用变长参数

   在 CUDA 编程中，核函数__global__不能使用变长参数，其缘故原由主要与并行盘算模型、线程独立性、内存访问模式以及性能优化相干。以下是一些关键点：
  

• 线程的独立性：CUDA 的核心计划思想是每个线程应当独立运行，变长参数的使用可能导致每个线程实验的代码路径不同，这粉碎了并行盘算的规则，使得步伐的实验变得不可猜测。每个线程的实验应该尽可能雷同且高效。
  
• 性能题目：使用变长参数会增长运行时的解析和内存分配开销，这可能导致盘算瓶颈，影响 CUDA 核函数的高效性。变长参数通常涉及动态内存分配、类型推导等操作，这些都是在编译时无法确定的，会增长 GPU 上并行盘算的额外负担。
  
• 内存访问冲突： 变长参数的动态性质可能导致线程间不同等的内存访问模式，从而引发内存访问冲突和性能降落。CUDA 要求内存访问尽量是局部和同等的，而变长参数增长了不确定性。
  
• 硬件资源的限制： GPU 盘算中的每个线程都有其固定的资源（如寄存器、内存等），变长参数的使用可能会导致这些资源的分配不均，影响到盘算资源的最大化利用。
  

  3.3 核函数为什么不能使用静态变量

   在 CUDA 编程中，核函数不能使用静态变量，缘故原由如下：
  

• 线程之间的资源共享题目：
  静态变量通常在步伐运行过程中只有一个实例，而核函数的每个线程都会独立实验。如果在核函数中使用静态变量，那么多个线程可能会并发访问同一份数据，这可能导致线程间的同步题目或数据竞争。CUDA 计划要求每个线程都应该独立，不能有共享的状态（除非通过显式的同步机制）。
  
• 线程的独立性：
  在 CUDA 中，每个线程都应该是独立的，能够并行实验而不依赖于其他线程的实验状态。静态变量的生命周期通常会超过多个函数调用，因此可能会导致线程在实验过程中产生意外的相互依赖，从而降低并行实验的效率。
  
• 内存管理题目：
  静态变量的内存分配通常是在步伐启动时完成的，而 CUDA 步伐中，内存管理依赖于 GPU 的内存模型，且不同的线程可能需要独立的数据。如果使用静态变量，会导致内存分配和访问的不同等，进而影响内存的使用效率和步伐的稳定性。
  

  3.4 核函数为什么不能使用函数指针

   在 CUDA 中，核函数不能使用函数指针，这一限制主要源于以下几个缘故原由：
  

• 线程实验路径的不可猜测性：
  核函数的每个线程都是独立实验的，而且实验路径应该是固定和同等的。函数指针通常会导致实验路径的动态决定，使得每个线程的实验路径可能不同。这种不确定性会影响到 GPU 资源的调度和优化，特别是在高并发的情况下。
  
• 编译时优化题目：
  使用函数指针会使得编译器无法在编译时确定函数的调用目的，导致无法进行高效的优化。CUDA 核函数通常依赖于编译时的优化（如内联、循环展开等），而使用函数指针会增长编译时的复杂性，影响最终的代码生成和性能。
  
• 硬件实验模型的限制：
  在 GPU 上，线程的实验通常是基于硬件的实验模型进行的，线程之间的实验应当是简单、快速且高效的。函数指针的使用可能导致线程实验时需要额外的跳转和盘算，增长了硬件实验路径的复杂性，进而影响性能。
  
• 内存管理题目：
  函数指针的使用可能会导致内存访问的不同等性，特别是在多个线程中。由于函数指针可能指向不同的代码路径，可能会出现不同线程间对内存访问模式的竞争，这不仅会降低步伐效率，还可能导致并发访问的错误。
  

  3.5 怎样理解核函数具有异步性

   核函数的异步性 是 CUDA 编程中的一个重要概念。它指的是在 GPU 上启动的核函数的实验与主机（CPU）步伐的实验是异步的，也就是说，核函数的调用并不会立即阻塞主机步伐的实验。以下是异步性的一些关键点：
  

• 核函数调用的异步实验：
  当你调用一个核函数（例如 kernel<<<blocks, threads>>>()），这并不会立即阻塞主机步伐。相反，主机步伐会继承实验下去，直到显式的同步操作发生，大概核函数实验完成后，CUDA 运行时会同步主机和设备的实验。
  
• 显式同步：
  核函数的异步性要求在需要时显式地进行同步。例如，使用 cudaDeviceSynchronize() 可以确保主机步伐期待 GPU 上的核函数实验完毕。没有这种同步机制，主机步伐可能会在核函数实验完成之前就退出，导致错误或不同等的效果。
  
• CUDA 内核调度与实验：
  GPU 会根据资源的可用性调度核函数的实验。这意味着，核函数的实验可能在后台启动，而主机步伐不会期待它完成。这种异步模型最大化了 CPU 和 GPU 的并行性，使得主机步伐可以在核函数实验期间实验其他任务，提高团体性能。
  
• 硬件调度的异步性：
  GPU 中的多个 SM（流式多处理器）能够同时实验多个核函数调用，从而实现并行盘算。核函数的实验在硬件级别是异步的，不同线程和块之间可能会并行实验。GPU 内部的线程调度与主机的调度相互独立，增长了灵活性和高效性。
  
• 性能优化：
  异步性使得 GPU 可以在期待盘算完成时进行其他工作，如处理更多的线程或实验其他核函数。在高性能盘算场景中，这种异步特性有助于提升 GPU 的利用率和团体性能。通过合理安排同步操作，步伐员可以最大化 CPU 和 GPU 的并行盘算。
  

  本文参考：
https://www.bilibili.com/video/BV1sM4y1x7of/?spm_id_from=333.788.videopod.episodes&vd_source=cf0b4c9c919d381324e8f3466e714d7a&p=4

免责声明：如果侵犯了您的权益，请联系站长，我们会及时删除侵权内容，谢谢合作！更多信息从访问主页：qidao123.com:ToB企服之家，中国第一个企服评测及商务社交产业平台。