专注于人工智能领域的小何尚

内核缓存有 gen 和 bin 文件夹，分别包含生成的 C++/CUDA 代码和编译的二进制文件。对于复杂的应用程序，可能很难理解导致错误的操作顺序。如果遇到错误，它将引发异常，这通常有助于缩小有问题的内核范围。通常，最好的调试方法之一是简单地打印内核中的值。此设置可确保正确生成行号和调试符号。启动 Python 进程后，应附加调试器，并将断点插入到生成的代码中。Warp 将在调试构建配置中执行边界检查，以确保所有数组访问都位于定义的形状内。，了解如何调试在 CPU 上运行的 Warp 内核代码的示例。

Warp 可以通过标准接口协议与其他基于 Python 的框架（例如 NumPy）进行互操作。

Warp 支持编写通用内核和函数，它们充当可以使用不同具体类型实例化的模板。这使您可以编写一次代码并以多种数据类型重用它。

在使用多个流的程序中，这比 wp.synchronize_device() 提供了对同步行为更细粒度的控制，wp.synchronize_device() 会同步设备上的所有流。问题是内核是在流 s 上启动的，这与用于创建数组 a 的流不同。该数组是在设备 cuda:0 的当前流上分配和初始化的，这意味着当流 s 开始执行使用该数组的内核时，该数组可能尚未准备好。事件可用于同步同一设备甚至不同 CUDA 设备上的流，因此您可以编排完全在可用 GPU 上执行的非常复杂的多流和多设备工作负载。

Warp 0.14.0 添加了对 CUDA 数组的流排序内存池分配器的支持。从 Warp 0.15.0 开始，这些分配器在所有支持它们的 CUDA 设备上默认启用。“流顺序内存池分配器”很拗口，所以让我们一次一点地解开它。上面的每个调用都会分配一块足够大的设备内存来容纳数组，并可以选择使用指定的值初始化内容。wp.empty() 是唯一一个不以任何方式初始化内容的函数，它只是分配内存。内存池分配器从更大的保留内存池中获取一块内存，这通常比向操作系统请求全新的存储块要快。

Warp 旨在与任意 CUDA 上下文配合使用，因此它可以轻松集成到不同的工作流程中。基于 CUDA Runtime API 构建的应用程序针对每个设备的主要上下文。运行时 API 在幕后隐藏了 CUDA 上下文管理。在 Warp 中，设备“cuda:i”代表设备 i 的主要上下文，它与 CUDA 运行时 API 一致。基于 CUDA 驱动程序 API 构建的应用程序直接使用 CUDA 上下文，并且可以在任何设备上创建自定义 CUDA 上下文。

由于 Warp 内核被编译为本机 C++/CUDA 代码，因此所有函数输入参数都必须是静态类型的。这将自动与 GPU 同步，以确保所有未完成的工作已完成，并将数组复制回 CPU 内存，并在其中传递给 NumPy。默认情况下，从外部数据（例如：NumPy、列表、元组）初始化的 Warp 数组将为指定设备创建数据副本到新内存。属于 Python 模块的所有内核都会在运行时编译为动态库和 PTX，然后在应用程序重新启动之间缓存结果，以实现快速启动。用户函数可以在同一模块内的内核中自由调用，并接受数组作为输入。

Warp 是一个用于编写高性能模拟和图形代码的 Python 框架。Warp 采用常规 Python 函数，JIT 将它们编译为可以在 CPU 或 GPU 上运行的高效内核代码。Warp 专为空间计算而设计，并附带一组丰富的原语，可以轻松编写物理模拟、感知、机器人和几何处理程序。此外，Warp 内核是可微分的，可以用作 PyTorch 和 JAX 等框架的机器学习管道的一部分。

本文介绍如何在win10系统中搭建CUDA + VS2019的开发环境CUDA是目前做人工智能, 深度学习等方向的必备工具库. 由CUDA衍生出的加速工具很多, 如: cuDNN, TensorRT, cuBLAS等HPC加速库, 或者涉及最新的元宇宙概念中的Omniverse等.

1.CUDA简介1.1 我们为什么要使用GPUGPU（Graphics Processing Unit）在相同的价格和功率范围内，比CPU提供更高的指令吞吐量和内存带宽。许多应用程序利用这些更高的能力，在GPU上比在CPU上运行得更快(参见GPU应用程序)。其他计算设备，如FPGA，也非常节能，但提供的编程灵活性要比GPU少得多。GPU和CPU在功能上的差异是因为它们的设计目标不同。虽然 CPU 旨在以尽可能快的速度执行一系列称为线程的操作，并且可以并行执行数十个这样的线程。但GPU却能并行执行成千上

2.编程模型本章通过概述CUDA编程模型是如何在c++中公开的，来介绍CUDA的主要概念。编程接口中给出了对 CUDA C++ 的广泛描述。本章和下一章中使用的向量加法示例的完整代码可以在 vectorAdd CUDA示例中找到。2.1 内核CUDA C++ 通过允许程序员定义称为kernel的 C++ 函数来扩展 C++，当调用内核时，由 N 个不同的 CUDA 线程并行执行 N 次，而不是像常规 C++ 函数那样只执行一次。使用 __global__ 声明说明符定义内核，并使用新的 <

第三章编程接口CUDA C++ 为熟悉 C++ 编程语言的用户提供了一种简单的途径，可以轻松编写由设备执行的程序。它由c++语言的最小扩展集和运行时库组成。编程模型中引入了核心语言扩展。它们允许程序员将内核定义为 C++ 函数，并在每次调用函数时使用一些新语法来指定网格和块的维度。所有扩展的完整描述可以在 C++ 语言扩展中找到。任何包含这些扩展名的源文件都必须使用 nvcc 进行编译，如使用NVCC编译中所述。运行时在 CUDA Runtime 中引入。它提供了在主机上执行的 C 和 C++ 函数

第四章 硬件实现NVIDIA GPU 架构围绕可扩展的多线程流式多处理器 (SM: Streaming Multiprocessors) 阵列构建。当主机 CPU 上的 CUDA 程序调用内核网格时，网格的块被枚举并分发到具有可用执行能力的多处理器。一个线程块的线程在一个SM上并发执行，多个线程块可以在一个SM上并发执行。当线程块终止时，新块在空出的SM上启动。SM旨在同时执行数百个线程。为了管理如此大量的线程，它采用了一种称为 SIMT（Single-Instruction, Multiple-Thr

第五章 性能指南5.1 整体性能优化策略性能优化围绕四个基本策略：最大化并行执行以实现最大利用率；优化内存使用，实现最大内存吞吐量；优化指令使用，实现最大指令吞吐量；尽量减少内存抖动。哪些策略将为应用程序的特定部分产生最佳性能增益取决于该部分的性能限值； 例如，优化主要受内存访问限制的内核的指令使用不会产生任何显着的性能提升。 因此，应该通过测量和监控性能限制来不断地指导优化工作，例如使用 CUDA 分析器。 此外，将特定内核的浮点运算吞吐量或内存吞吐量（以更有意义的为准）与设备的相应峰值

附录A 支持GPU设备列表https://developer.nvidia.com/cuda-gpus 列出了所有支持 CUDA 的设备及其计算能力。可以使用运行时查询计算能力、多处理器数量、时钟频率、设备内存总量和其他属性（参见参考手册）。附录B 对C++扩展的详细描述B.1 函数执行空间说明符函数执行空间说明符表示函数是在主机上执行还是在设备上执行，以及它是可从主机调用还是从设备调用。B.1.1 __global____global__ 执行空间说明符将函数声明为内核。 它的功能是：在

附录C 协作组C.1. IntroductionCooperative Groups 是 CUDA 9 中引入的 CUDA 编程模型的扩展，用于组织通信线程组。协作组允许开发人员表达线程通信的粒度，帮助他们表达更丰富、更有效的并行分解。从历史上看，CUDA 编程模型为同步协作线程提供了一个单一、简单的构造：线程块的所有线程之间的屏障，如使用 __syncthreads() 内部函数实现的那样。但是，程序员希望以其他粒度定义和同步线程组，以“集体”组范围功能接口的形式实现更高的性能、设计灵活性和软件重用

附录D-CUDA的动态并行D.1. IntroductionD.1.1. Overview***Dynamic Parallelism***是 CUDA 编程模型的扩展，使 CUDA 内核能够直接在 GPU 上创建新工作并与新工作同步。在程序中需要的任何位置动态创建并行性提供了令人兴奋的新功能。直接从 GPU 创建工作的能力可以减少在主机和设备之间传输执行控制和数据的需要，因为现在可以通过在设备上执行的线程在运行时做出启动配置决策。此外，可以在运行时在内核内内联生成依赖于数据的并行工作，动态利用 G

附录E虚拟内存管理E.1. Introduction虚拟内存管理 API 为应用程序提供了一种直接管理统一虚拟地址空间的方法，该空间由 CUDA 提供，用于将物理内存映射到 GPU 可访问的虚拟地址。在 CUDA 10.2 中引入的这些 API 还提供了一种与其他进程和图形 API（如 OpenGL 和 Vulkan）进行互操作的新方法，并提供了用户可以调整以适应其应用程序的更新内存属性。从历史上看，CUDA 编程模型中的内存分配调用（例如 cudaMalloc）返回了一个指向 GPU 内存的内存地址

附录F 流序内存分配F.1. Introduction使用 cudaMalloc 和 cudaFree 管理内存分配会导致 GPU 在所有正在执行的 CUDA 流之间进行同步。 Stream Order Memory Allocator 使应用程序能够通过启动到 CUDA 流中的其他工作（例如内核启动和异步拷贝）来对内存分配和释放进行排序。这通过利用流排序语义来重用内存分配来改进应用程序内存使用。分配器还允许应用程序控制分配器的内存缓存行为。当设置了适当的释放阈值时，缓存行为允许分配器在应用程序表明它愿

附录G 图内存节点G.1. Introduction图内存节点允许图创建和拥有内存分配功能。图内存节点具有 GPU 有序生命周期语义，它指示何时允许在设备上访问内存。这些 GPU 有序生命周期语义支持驱动程序管理的内存重用，并与流序分配 API cudaMallocAsync 和 cudaFreeAsync 相匹配，这可能在创建图形时被捕获。图分配在图的生命周期内具有固定的地址，包括重复的实例化和启动。这允许图中的其他操作直接引用内存，而无需更新图，即使 CUDA 更改了后备物理内存也是如此。在一个图

附录H 数学方法参考手册列出了设备代码中支持的 C/C++ 标准库数学函数的所有函数及其描述，以及所有内部函数（仅在设备代码中支持）。本附录在适用时提供了其中一些功能的准确性信息。它使用 ULP 进行量化。有关最后位置单元 (ULP: Unit in the Last Place, 上面是直译的,这里可以理解为最小精度单元) 定义的更多信息，请参阅 Jean-Michel Muller’s paper On the definition of ulp(x), RR-5504, LIP RR-2005-0

附录I C++ 语言支持如使用 NVCC 编译中所述，使用 nvcc 编译的 CUDA 源文件可以包含主机代码和设备代码的混合。 CUDA 前端编译器旨在模拟主机编译器对 C++ 输入代码的行为。 输入源代码根据 C++ ISO/IEC 14882:2003、C++ ISO/IEC 14882:2011、C++ ISO/IEC 14882:2014 或 C++ ISO/IEC 14882:2017 规范进行处理，CUDA 前端编译器旨在模拟 任何主机编译器与 ISO 规范的差异。 此外，支持的语言使用本文

附录J 纹理获取本附录给出了用于计算 Texture Functions 的纹理函数返回值的公式，具体取决于纹理引用的各种属性（请参阅纹理和表面内存）。绑定到纹理引用的纹理表示为一个数组 T一维纹理的 N 个texels，二维纹理的 N x M texels，三维纹理的 N x M x L texels。它是使用非归一化纹理坐标 x、y 和 z 或归一化纹理坐标 x/N、y/M 和 z/L 获取的，如纹理内存中所述。 在本附录中，假定坐标在有效范围内。 纹理内存解释了如何根据寻址模式将超出范

附录K CUDA计算能力计算设备的一般规格和功能取决于其计算能力（请参阅计算能力）。下面的表格中 显示了与当前支持的每种计算能力相关的特性和技术规格。浮点标准审查是否符合 IEEE 浮点标准。Compute Capability 3.x、Compute Capability 5.x、Compute Capability 6.x、Compute Capability 7.x 和 Compute Capability 8.x 部分提供了有关计算能力 3.x、5.x、6 的设备架构的更多详细信息 .x、7

附录L CUDA底层驱动API本附录假定您了解 CUDA 运行时中描述的概念。驱动程序 API 在 cuda 动态库（cuda.dll 或 cuda.so）中实现，该库在安装设备驱动程序期间复制到系统上。 它的所有入口点都以 cu 为前缀。它是一个基于句柄的命令式 API：大多数对象都由不透明的句柄引用，这些句柄可以指定给函数来操作对象。驱动程序 API 中可用的对象汇总在下表中。 Tab

附录M_CUDA环境变量下表列出了 CUDA 环境变量。 与多进程服务相关的环境变量记录在 GPU 部署和管理指南的多进程服务部分。 Table 18. CUDA Environment Variables Var

附录N CUDA的统一内存N.1. Unified Memory Introduction统一内存是 CUDA 编程模型的一个组件，在 CUDA 6.0 中首次引入，它定义了一个托管内存空间，在该空间中所有处理器都可以看到具有公共地址空间的单个连贯内存映像。注意：处理器是指任何具有专用 MMU 的独立执行单元。这包括任何类型和架构的 CPU 和 GPU。底层系统管理 CUDA 程序中的数据访问和位置，无需显式内存复制调用。这在两个主要方面有利于 GPU 编程：通过统一系统中所有 GPU 和 CP