【C++】CUDA核函数

🕗 发布于 2024-12-09 22:42 c++ 算法 开发语言

文章目录

1. 从C++到CUDA编程
2. 核函数
3. 问题与思考

1. 从C++到CUDA编程

cuda中的Hello World程序

nvcc：

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

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

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

注意：在cuda编程中，不能使用cout进行输出，只能使用printf

2. 核函数

核函数简介：

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

定义核函数：

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

核函数注意事项：

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

核函数编写流程：

int main(){
主机代码
核函数调用
主机代码
return 0;
}

代码示例：

#include <stdio.h>

__global__ void hello_from_gpu(){
    printf("Hello World from GPU \n");
}

int main(){
    hello_from_gpu<<<1,6>>>();
    // 同步主机与设备，促使缓冲区刷新，打印信息到终端。该函数会等待GPU设备执行完毕
    cudaDeviceSynchronize();
    return 0;
}

代码解析：

1.核函数hello_from_gpu

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

2.核函数启动

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

3.同步

使用cudaDeviceSynchronize();来同步主机与设备，确保 GPU 上的所有线程执行完毕后，主机程序才继续执行

代码输出：

Hello World from GPU 
Hello World from GPU 
Hello World from GPU 
Hello World from GPU 
Hello World from GPU 
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

原文地址：https://blog.csdn.net/qq_42761751/article/details/144297840

免责声明：本站文章内容转载自网络资源，如本站内容侵犯了原著者的合法权益，可联系本站删除。更多内容请关注自学内容网（zxcms.com）！

上一篇：如果用Python写爬虫，具体怎么实现随机请求间隔呢？
下一篇：【JAVA练习】力扣860.柠檬水找零

.NET(C#) 如何配置用户首选项及保存用户设置
.NET(C#) 如何配置用户首选项及保存用户设置
阅读更多2024-12-14
【最新】北大数字普惠金融指数数据集-省市县（2011-2023年）
郭峰,王靖一,王芳,孔涛,张勋,程志云.测度中国数字普惠金融发展:指数编制与空间特征[J].经济学(季刊),2020,19(04):1401-1418.时间跨度：省级和城市级指数时间跨度为2011-2
阅读更多2024-12-14
GESP202412 四级【Recamán】题解（AC）
a11ak−1−kkakak−1−kak−1k小杨想知道 Recamán 数列的前n项从小到大排序后的结果。手动计算非常困难，小杨希望你能帮他解决这个问题。
阅读更多2024-12-14
IDEA遇到EasyConnect中的网络资源无法访问的问题
版权声明：本文为博主原创文章，遵循 CC 4.0 BY-SA 版权协议，转载请附上原文出处链接和本声明。原文链接：https://blog.csdn.net/wanshanyu_/article/de
阅读更多2024-12-14
双目摄像头标定方法
此时已经完成标定，左下角为反投影误差，右边为外参可视化。将双目左右目拍的图像上传（左右目最好不少于20张）此时回到主页面，即可看到成功导出。把这些误差大的删除即可。
阅读更多2024-12-14
Servlet、omcat服务器架构与工作原理
Servlet是运行在服务器端的Java程序，它的主要职责之一是接收并处理来自客户端（如浏览器）的HTTP请求。当客户端发送一个请求到服务器时，Servlet可以解析请求中的信息，例如请求的URL路径
阅读更多2024-12-14
Vue生命周期钩子函数：深入解析与实践
作为高级Vue前端开发人员，对Vue组件的生命周期钩子函数有着深刻的理解是至关重要的。生命周期钩子函数是指在Vue组件的创建、更新、销毁等过程中，Vue自动调用的一系列方法。通过这些钩子函数，我们可以
阅读更多2024-12-14
安卓开发--使用android studio发布APP
app发布
阅读更多2024-12-14
数据结构与算法学习笔记----拓扑排序
@ author: 明月清了个风。
阅读更多2024-12-14
python 将数据保存到现有的Excel文件的新工作表
out_file = ‘query.xlsx’df1 = pd.DataFrame(out_data)若直接写入：df1.to_excel(out_file, index=False, sheet_n
阅读更多2024-12-14