转载请注明出处：小锋学长生活大爆炸[xfxuezhagn.cn]

如果本文帮助到了你，欢迎[点赞、收藏、关注]哦~

---

内容较多，可通过 左侧或右侧 的 目录 进行跳转。

---

        CUDA 是“Compute Unified Device Architecture (计算统一设备架构)”的首字母缩写。CUDA 是一种用于并行计算的 NVIDIA 架构。使用图形处理器也可以提高 PC 的计算能力。

Samples list

0. Introduction

        这些示例展示了 CUDA 编程的各种基本和高级技术，从简单的算术运算到复杂的并行计算和优化策略，为用户提供了丰富的学习和实践资源。

        介绍。此部分包含针对初学者的基本 CUDA 示例，展示了使用 CUDA 和 CUDA 运行时 API 的关键概念。目的是帮助新手快速理解和入门 CUDA 编程。

asyncAPI

        此示例展示了如何使用 CUDA 事件进行 GPU 计时以及重叠 CPU 和 GPU 的执行。在 CUDA 调用流中插入事件。由于 CUDA 流调用是异步的，CPU 可以在 GPU 执行期间进行计算（包括主机和设备之间的 DMA 内存复制）。CPU 可以查询 CUDA 事件，以确定 GPU 是否完成任务。

c++11_cuda

        此示例展示了 CUDA 中对 C++11 特性的支持。它扫描一个输入文本文件并打印 x、y、z、w 字符的出现次数。

clock

        这个示例展示了如何使用 clock 函数准确测量一个内核中线程块的性能。

clock_nvrtc

        这个示例展示了如何使用 libNVRTC 中的 clock 函数来准确测量一个内核中线程块的性能。

concurrentKernels

        此示例展示了如何使用 CUDA 流来同时在 GPU 设备上执行多个内核。还展示了如何使用新的 cudaStreamWaitEvent 函数在 CUDA 流之间引入依赖关系。

cppIntegration

        这个示例展示了如何将 CUDA 集成到现有的 C++ 应用程序中，即在主机端的 CUDA 入口点只是从 C++ 代码调用的一个函数，并且只有包含该函数的文件使用 nvcc 编译。它还展示了如何在 C++ 中使用向量类型。

cppOverload

        这个示例展示了如何在 GPU 上使用 C++ 函数重载。

cudaOpenMP

        这个示例展示了如何使用 OpenMP API 编写多 GPU 应用程序。

fp16ScalarProduct

        计算两个 FP16 数字向量的标量积。

matrixMul

        这个示例实现了矩阵乘法，与编程指南第 6 章完全相同。它是为了清晰地说明各种 CUDA 编程原则，而不是为了提供最通用的高性能矩阵乘法内核。为了展示 GPU 的矩阵乘法性能，该示例还展示了如何使用新的 CUDA 4.0 接口 CUBLAS 实现高性能矩阵乘法。

matrixMul_nvrtc

        这个示例实现了矩阵乘法，与编程指南第 6 章完全相同。它是为了清晰地说明各种 CUDA 编程原则，而不是为了提供最通用的高性能矩阵乘法内核。为了展示 GPU 的矩阵乘法性能，该示例还展示了如何使用新的 CUDA 4.0 接口 CUBLAS 实现高性能矩阵乘法。

matrixMulDrv

        这个示例实现了矩阵乘法，并使用了新的 CUDA 4.0 内核启动驱动 API。它是为了清晰地说明各种 CUDA 编程原则，而不是为了提供最通用的高性能矩阵乘法内核。CUBLAS 提供高性能的矩阵乘法。

matrixMulDynlinkJIT

        这个示例重新探讨了使用 CUDA 驱动 API 的矩阵乘法。它展示了如何在运行时链接到 CUDA 驱动程序以及如何使用 PTX 代码进行 JIT（即时）编译。它是为了清晰地说明各种 CUDA 编程原则，而不是为了提供最通用的高性能矩阵乘法内核。CUBLAS 提供高性能的矩阵乘法。

mergeSort

        这个示例实现了归并排序（也称为 Batcher 排序），属于排序网络类算法。虽然在大序列上一般效率较低，但在对短至中等大小的（键，值）数组对进行排序时，可能是优选算法。参考 H. W. Lang 的优秀教程：http://www.iti.fh-flensburg.de/lang/algorithmen/sortieren/networks/indexen.htm

simpleAssert

        这个 CUDA 运行时 API 示例是一个非常基础的示例，展示了如何在设备代码中使用 assert 函数。需要计算能力 2.0。

simpleAssert_nvrtc

        这个 CUDA 运行时 API 示例是一个非常基础的示例，展示了如何在设备代码中使用 assert 函数。需要计算能力 2.0。

simpleAtomicIntrinsics

        全局内存原子指令的简单演示。

simpleAtomicIntrinsics_nvrtc

        全局内存原子指令的简单演示。此示例使用 NVRTC 进行运行时编译。

simpleAttributes

        这个 CUDA 运行时 API 示例是一个非常基础的示例，展示了如何使用影响 L2 局部性的流属性。由于使用 L2 访问策略窗口带来的性能提升只能在计算能力 8.0 或更高的设备上注意到。

simpleAWBarrier

        到达等待屏障的简单演示。

simpleCallback

        这个示例实现了多线程异构计算工作负载，使用 CUDA 5.0 引入的 CUDA 流和事件的新 CPU 回调。

simpleCooperativeGroups

        这个示例是一个简单的代码，展示了在线程块内使用协作组的基本用法。

simpleCubemapTexture

        简单示例，展示了如何使用 CUDA C 中的新 CUDA 4.1 特性来支持立方图纹理。

simpleCUDA2GL

        这个示例展示了如何使用最有效的方法将 CUDA 图像复制回 OpenGL。

simpleDrvRuntime

        一个简单的示例，展示了 CUDA 驱动程序和运行时 API 如何协同工作，加载向量加法内核的 cuda fatbinary 并执行向量加法。

simpleHyperQ

        这个示例展示了如何使用 CUDA 流在提供 HyperQ（SM 3.5）的设备上同时执行多个内核。在没有 HyperQ（SM 2.0 和 SM 3.0）的设备上，将最多同时运行两个内核。

simpleIPC

        这个 CUDA 运行时 API 示例是一个非常基础的示例，展示了每个 GPU 一个进程的进程间通信。需要计算能力 3.0 或更高的设备和 Linux 操作系统，或者启用了 TCC 的 Windows 操作系统。

simpleLayeredTexture

        简单示例，展示了如何使用 CUDA C 中的新 CUDA 4.0 特性来支持分层纹理。

simpleMPI

        简单示例，展示了如何结合使用 MPI 和 CUDA。

simpleMultiCopy

        在计算能力 1.1 的 GPU 中，支持使用一个 memcopy 重叠计算。从主机系统的。对于具有计算能力 2.0 的 Quadro 和 Tesla GPU，可能以全速进行第二个重叠复制操作（PCI-e 是对称的）。此示例展示了如何使用 CUDA 流实现内核执行与设备之间的数据复制的重叠。

simpleMultiGPU

        这个应用程序展示了如何使用新的 CUDA 4.0 API 进行 CUDA 上下文管理和多线程访问，以在多 GPU 上运行 CUDA 内核。

simpleOccupancy

        这个示例展示了如何使用 CUDA 占用率计算器和基于占用率的启动配置 API，通过启动配置器启动内核，并测量与手动配置启动的利用率差异。

simpleP2P

        这个应用程序展示了 CUDA API 如何支持多 GPU 之间的点对点（P2P）复制、点对点（P2P）寻址和统一虚拟内存寻址（UVA）。一般来说，P2P 支持在两个相同的 GPU 之间，某些例外如一些 Tesla 和 Quadro GPU。

simplePitchLinearTexture

        使用线性纹理示例。

simplePrintf

        这个基础的 CUDA 运行时 API 示例展示了如何在设备代码中使用 printf 函数。

simpleSeparateCompilation

        这个示例展示了 CUDA 5.0 特性，即创建 GPU 设备静态库并在另一个 CUDA 内核中使用它的能力。该示例展示了如何将 GPU 设备函数（来自 GPU 设备静态库）作为函数指针传递以供调用。此示例需要计算能力 2.0 或更高的设备。

simpleStreams

        这个示例使用 CUDA 流重叠内核执行与主机和 GPU 设备之间的内存复制。此示例使用了一种新的 CUDA 4.0 特性，该特性支持固定通用主机内存。需要计算能力 2.0 或更高。

simpleSurfaceWrite

        使用 2D 表面引用（写入纹理）的简单示例。

simpleTemplates

        这个示例是模板项目的模板化版本。它还展示了如何正确模板化动态分配的共享内存数组。

simpleTemplates_nvrtc

        这个示例是模板项目的模板化版本。它还展示了如何正确模板化动态分配的共享内存数组。

simpleTexture

        使用纹理的简单示例。

simpleTexture3D

        使用 3D 纹理的简单示例。

simpleTextureDrv

        使用纹理的简单示例。这个示例使用了新的 CUDA 4.0 内核启动驱动 API。

simpleVoteIntrinsics

        一个简单程序，展示了如何在 CUDA 内核中使用投票（__any_sync, __all_sync）内在函数。

simpleVoteIntrinsics_nvrtc

        一个简单程序，展示了如何在 CUDA 内核中使用投票（any, all）内在函数，使用 NVRTC API 进行运行时编译。需要计算能力 2.0 或更高的设备。

simpleZeroCopy

        这个示例展示了如何使用零内存复制，内核可以直接读取和写入固定系统内存。

systemWideAtomics

        一个系统范围原子指令的简单演示。

template

        一个简单的模板项目，可以用作创建新 CUDA 项目的起点。

UnifiedMemoryStreams

        这个示例展示了在单 GPU 上使用 OpenMP 和流与统一内存的示例。

vectorAdd

        这个 CUDA 运行时 API 示例是一个非常基础的示例，实现了逐元素向量加法。与编程指南第 3 章的示例相同，并添加了一些错误检查。

vectorAdd_nvrtc

        这个 CUDA 驱动 API 示例使用 NVRTC 进行向量加法内核的运行时编译。向量加法内核与编程指南第 3 章的示例相同。

vectorAddDrv

        这个向量加法示例是逐元素实现的基础示例。与编程指南第 3 章的示例相同，并添加了一些错误检查。此示例还使用了新的 CUDA 4.0 内核启动驱动 API。

vectorAddMMAP

        这个示例用 cuMemMap-ed 分配替换了 vectorAddDrv 示例中的设备分配。此示例展示了 cuMemMap API 允许用户指定其内存的物理属性，同时保留其访问的连续性，从而不需要更改程序结构。

1. Utilities

        这些实用工具示例为用户提供了方便的方法来测量和查询系统中 CUDA 设备的性能和属性，有助于优化 CUDA 应用程序的性能和资源利用。

        实用工具。这一部分的示例展示了如何查询设备的能力以及如何测量 GPU 和 CPU 之间的带宽。这些实用工具对于了解硬件性能和优化数据传输非常有用。

bandwidthTest

        这是一个简单的测试程序，用于测量 GPU 的内存复制带宽和 PCI-e 上的内存复制带宽。这个测试应用程序能够测量设备到设备的复制带宽、页面内存和页锁定内存的主机到设备复制带宽，以及页面内存和页锁定内存的设备到主机复制带宽。

deviceQuery

        这个示例列举了系统中存在的 CUDA 设备的属性。它可以帮助用户了解系统中每个 CUDA 设备的详细信息，如设备名称、计算能力、可用内存等。

deviceQueryDrv

        这个示例使用 CUDA 驱动程序 API 调用列举系统中 CUDA 设备的属性。它与 deviceQuery 示例类似，但使用的是驱动程序 API 而不是运行时 API。

topologyQuery

        这是一个简单的示例，展示了如何查询具有多个 GPU 的系统的拓扑结构。它可以帮助用户了解 GPU 之间以及 GPU 和其他系统组件之间的连接和关系，从而优化数据传输和计算任务的分配。

2. Concepts and Techniques

        概念和技术。此部分的示例展示了与 CUDA 相关的概念以及解决常见问题的方法。例如，如何有效地管理内存、优化线程调度、处理并行计算中的常见挑战等。

boxFilter

        使用 CUDA 和 OpenGL 渲染的快速图像方框滤波器。

convolutionSeparable

        这个示例实现了一个带有高斯核的 2D 信号的可分离卷积滤波器。

convolutionTexture

        基于纹理的带有高斯核的 2D 可分离卷积实现。用于与 convolutionSeparable 的性能比较。

cuHook

        这个示例展示了如何构建和使用 CUDA 拦截库。库需要通过 LD_PRELOAD 加载，例如：LD_PRELOAD=<full_path>/libcuhook.so.1 ./cuHook。

dct8x8

        此示例展示了如何使用 CUDA 对 8x8 像素块执行离散余弦变换（DCT）：一个基于定义的简单实现和许多库中使用的更传统的方法。与在片段着色器中实现 DCT 相比，CUDA 允许更简单和更高效的实现。

EGLStream_CUDA_CrossGPU

        展示了 CUDA 和 EGL 流的互操作性，其中消费者的 EGL 流在一个 GPU 上，生产者在另一个 GPU 上，并且消费者和生产者是不同的进程。

EGLStream_CUDA_Interop

        展示了 CUDA 和 EGL 流之间的数据交换。

EGLSync_CUDAEvent_Interop

        展示了 CUDA 事件与 EGL 同步/EGL 图像之间的互操作性，通过该互操作性，可以在 GPU 上实现 GL-EGL-CUDA 操作的同步，而不是阻塞 CPU 进行同步。

eigenvalues

        计算全部或部分特征值是线性代数、统计学、物理学和许多其他领域的重要问题。这个示例展示了用 CUDA 实现的并行二分算法，用于计算任意大小的对称三对角矩阵的所有特征值。

FunctionPointers

        这个示例展示了如何使用函数指针并实现 8 位单色图像的 Sobel 边缘检测滤波器。

histogram

        这个示例展示了 64-bin 和 256-bin 直方图的高效实现。

imageDenoising

        这个示例展示了两种自适应图像去噪技术：KNN 和 NLM，基于纹理元素之间的几何和颜色距离计算。虽然这两种技术在 DirectX SDK 中使用着色器实现，但该示例还实现了利用共享内存的更快速变体，以与 DirectX 对应物进行比较。

inlinePTX

        一个简单的测试应用程序，展示了 CUDA 4.0 新增功能，将 PTX 嵌入到 CUDA 内核中。

inlinePTX_nvrtc

        一个简单的测试应用程序，展示了 CUDA 4.0 新增功能，将 PTX 嵌入到 CUDA 内核中。

interval

        区间算术运算符示例。使用了各种 C++ 特性（模板和递归）。递归模式需要计算能力 SM 2.0。

MC_EstimatePiInlineP

        这个示例使用蒙特卡罗模拟来估算 Pi（使用内联 PRNG）。这个示例还使用了 NVIDIA CURAND 库。

MC_EstimatePiInlineQ

        这个示例使用蒙特卡罗模拟来估算 Pi（使用内联 QRNG）。这个示例还使用了 NVIDIA CURAND 库。

MC_EstimatePiP

        这个示例使用蒙特卡罗模拟来估算 Pi（使用批量 PRNG）。这个示例还使用了 NVIDIA CURAND 库。

MC_EstimatePiQ

        这个示例使用蒙特卡罗模拟来估算 Pi（使用批量 QRNG）。这个示例还使用了 NVIDIA CURAND 库。

MC_SingleAsianOptionP

        这个示例使用蒙特卡罗方法模拟单一亚洲期权，使用 NVIDIA CURAND 库。

particles

        这个示例使用 CUDA 来模拟和可视化大量粒子及其物理交互。添加 "-particles=" 到命令行将允许用户设置模拟粒子的数量。此示例使用原子操作或 Thrust 库中的快速基数排序实现了统一网格数据结构。

radixSortThrust

        这个示例展示了使用 Thrust 库的非常快速和高效的并行基数排序。包含的 RadixSort 类可以对键值对（浮点或无符号整数键）或仅对键进行排序。

reduction

        一个并行求和归约，计算大数组的值之和。这个示例展示了数据并行算法的重要优化策略，如使用共享内存、__shfl_down_sync、__reduce_add_sync 和 cooperative_groups reduce 进行归约。

reductionMultiBlockCG

        这个示例展示了使用多块协作组的单次归约。此示例需要具有计算抢占功能的计算能力 6.0 或更高版本的设备。

scalarProd

        这个示例计算给定输入向量对的标量积。

scan

        这个示例展示了并行前缀和（也称为“扫描”）的高效 CUDA 实现。给定一个数字数组，扫描计算一个新数组，其中每个元素是输入数组中该元素之前所有元素的和。

segmentationTreeThrust

        这个示例展示了图像分割树构建的方法。该方法基于 Boruvka 的 MST 算法。

shfl_scan

        这个示例展示了如何使用 shuffle 内在函数 __shfl_up_sync 在线程块中执行扫描操作。

sortingNetworks

        这个示例实现了 bitonic 排序和奇偶合并排序（也称为 Batcher 排序），属于排序网络类算法。虽然对于大序列来说效率较低，但对于短到中等大小的（键，值）数组对的排序，这可能是首选算法。参考 H. W. Lang 的优秀教程：http://www.iti.fh-flensburg.de/lang/algorithmen/sortieren/networks/indexen.htm

streamOrderedAllocation

        这个示例展示了使用 cudaMallocAsync 和 cudaMemPool 系列 API 在 GPU 上进行流有序内存分配。

streamOrderedAllocationIPC

        这个示例展示了使用 cudaMallocAsync 和 cudaMemPool 系列 API 分配的流有序内存的 IPC 池。

streamOrderedAllocationP2P

        这个示例展示了使用 cudaMallocAsync 和 cudaMemPool 系列 API 分配的流有序内存的对等访问。

threadFenceReduction

        这个示例展示了如何使用线程栅栏内在函数对值数组进行归约操作，以在单个内核中生成单个值（而不是像“reduction”CUDA 示例中那样调用两个或更多内核）。单次归约需要全局原子指令（计算能力 2.0 或更高）和 _threadfence() 内在函数（CUDA 2.2 或更高版本）。

threadMigration

        简单程序，展示了如何使用 CUDA 上下文管理 API，并使用新的 CUDA 4.0 参数传递和 CUDA 启动 API。CUDA 上下文可以分别创建并独立附加到不同的线程。

3. CUDA Features

        这些示例展示了 CUDA 的一些高级功能，如张量核心、动态并行、图形 API 等，帮助用户了解和利用这些功能来提高计算性能和效率。

        特性。这里的示例展示了 CUDA 的一些特性，比如协同组 (Cooperative Groups)、CUDA 动态并行 (CUDA Dynamic Parallelism)、CUDA 图 (CUDA Graphs) 等。通过这些示例，用户可以了解并使用这些高级特性来提升程序性能。

bf16TensorCoreGemm

        一个 CUDA 示例，展示了使用 CUDA 11 在 Ampere 芯片家族张量核心中引入的 Warp 矩阵乘法和累加 (WMMA) API 进行 __nv_bfloat16 (e8m7) GEMM 计算。该示例还使用了 CUDA 管道接口提供的异步复制，将全局内存数据复制到共享内存，从而提高内核性能并减少寄存器压力。

binaryPartitionCG

        这个示例是一个简单代码，展示了二进制分区协作组和线程块内的归约。

bindlessTexture

        这个示例展示了 cudaSurfaceObject、cudaTextureObject 和 CUDA 中的 MipMap 支持的使用。需要计算能力 SM 3.0 的 GPU 才能运行该示例。

cdpAdvancedQuicksort

        这个示例展示了使用 CUDA 动态并行实现的高级快速排序。该示例需要计算能力 3.5 或更高的设备。

cdpBezierTessellation

        这个示例展示了使用 CUDA 动态并行实现的贝塞尔曲线细分。该示例需要计算能力 3.5 或更高的设备。

cdpQuadtree

        这个示例展示了使用 CUDA 动态并行实现的四叉树。该示例需要计算能力 3.5 或更高的设备。

cdpSimplePrint

        这个示例展示了使用 CUDA 动态并行实现的简单 printf。该示例需要计算能力 3.5 或更高的设备。

cdpSimpleQuicksort

        这个示例展示了使用 CUDA 动态并行实现的简单快速排序。该示例需要计算能力 3.5 或更高的设备。

cudaCompressibleMemory

        这个示例展示了使用 cuMemMap API 的可压缩内存分配。

cudaTensorCoreGemm

        CUDA 示例展示了使用 CUDA 9 中引入的 Warp 矩阵乘法和累加 (WMMA) API 进行 GEMM 计算。该示例展示了使用 Volta 芯片家族中引入的张量核心进行更快速的矩阵运算。此外，它还展示了使用新的 CUDA 函数属性 cudaFuncAttributeMaxDynamicSharedMemorySize，允许应用程序预留比默认情况下更多的共享内存。

dmmaTensorCoreGemm

        CUDA 示例展示了使用 CUDA 11 中在 Ampere 芯片家族张量核心中引入的双精度 Warp 矩阵乘法和累加 (WMMA) API 进行双精度 GEMM 计算。该示例还使用了 CUDA 管道接口提供的异步复制，从全局内存到共享内存进行异步加载，从而提高内核性能并减少寄存器压力。此外，该示例还展示了如何使用协作组异步复制接口在组内执行全局内存到共享内存的异步加载。

globalToShmemAsyncCopy

        这个示例实现了矩阵乘法，使用了异步复制数据从全局内存到共享内存（计算能力 8.0 或更高）。还展示了用于同步的到达等待屏障。

graphMemoryFootprint

        这个示例展示了图形内存节点如何重新使用虚拟地址和物理内存。

graphMemoryNodes

        展示了使用图形 API 和流捕获 API 在 CUDA 图形中进行内存分配和释放。

immaTensorCoreGemm

        CUDA 示例展示了使用 CUDA 10 中引入的 Warp 矩阵乘法和累加 (WMMA) API 进行整数 GEMM 计算。该示例展示了使用 Volta 芯片家族中引入的张量核心进行更快速的矩阵运算。此外，它还展示了使用新的 CUDA 函数属性 cudaFuncAttributeMaxDynamicSharedMemorySize，允许应用程序预留比默认情况下更多的共享内存。

jacobiCudaGraphs

        使用 Jacobi 迭代法展示了实例化 CUDA 图形更新，使用 cudaGraphExecKernelNodeSetParams() 和 cudaGraphExecUpdate() 方法。

memMapIPCDrv

        这个 CUDA 驱动 API 示例是一个非常基础的示例，展示了使用 cuMemMap API 进行进程间通信，每个 GPU 一个进程进行计算。需要计算能力 3.0 或更高的设备和 Linux 操作系统，或者启用了 TCC 的 Windows 操作系统。

newdelete

        这个示例展示了通过设备 C++ new 和 delete 操作符以及 CUDA 4.0 提供的虚函数声明进行动态全局内存分配。

ptxjit

        这个示例使用驱动 API 从 PTX 代码进行即时 (JIT) 编译内核。此外，该示例展示了 CUDA 运行时和 CUDA 驱动 API 调用的无缝互操作性。对于 CUDA 5.5，该示例展示了如何使用 cuLink* 函数在运行时使用 CUDA 驱动程序链接 PTX 汇编。

simpleCudaGraphs

        展示了使用图形 API 和流捕获 API 创建、实例化和启动 CUDA 图形。

StreamPriorities

        这个示例展示了流优先级的基本使用。

tf32TensorCoreGemm

        一个 CUDA 示例，展示了使用 CUDA 11 在 Ampere 芯片家族张量核心中引入的 Warp 矩阵乘法和累加 (WMMA) API 进行 tf32 (e8m10) GEMM 计算。该示例还使用了 CUDA 管道接口提供的异步复制，从全局内存到共享内存进行异步加载，从而提高内核性能并减少寄存器压力。

warpAggregatedAtomicsCG

        这个示例展示了如何使用协作组 (CG) 执行 warp 聚合原子操作到单个或多个计数器，当许多线程原子地添加到单个或多个计数器时，这是一个有用的技术，可以提高性能。

graphConditionalNodes

        展示了从 CUDA 12.4 开始提供的 CUDA 图形条件节点的使用。

4. CUDA Libraries

        这些示例展示了如何使用 CUDA 平台库进行各种高级计算任务，从线性代数到图像处理和随机数生成，帮助用户了解和使用这些库来提高其 CUDA 应用程序的性能和功能。

        库。这一部分展示了如何使用 CUDA 平台提供的各种库，包括 NPP（NVIDIA Performance Primitives）、NVJPEG、NVGRAPH、cuBLAS、cuFFT、cuSPARSE、cuSOLVER 和 cuRAND 等。这些库提供了丰富的功能，用于图像处理、图形计算、线性代数、随机数生成等领域。

batchCUBLAS

        一个 CUDA 示例，展示了如何使用批量 CUBLAS API 调用来提高整体性能。

batchedLabelMarkersAndLabelCompressionNPP

        一个 NPP CUDA 示例，展示了如何使用基于联合查找 (UF) 算法的 NPP 标签标记生成和标签压缩功能，包括单个图像和批量图像版本。

boxFilterNPP

        一个 NPP CUDA 示例，展示了如何使用 NPP FilterBox 函数执行方框滤波。

cannyEdgeDetectorNPP

        一个 NPP CUDA 示例，展示了使用 nppiFilterCannyBorder_8u_C1R Canny 边缘检测图像滤波函数的推荐参数。该函数需要单通道 8 位灰度输入图像。可以通过首先调用 nppiColorToGray() 或 nppiRGBToGray() 从彩色图像生成灰度图像。Canny 边缘检测函数结合并改进了多步骤边缘检测图像的技术。

conjugateGradient

        这个示例使用 CUBLAS 和 CUSPARSE 库在 GPU 上实现共轭梯度求解器。

conjugateGradientCudaGraphs

        这个示例使用 CUBLAS 和 CUSPARSE 库调用，通过 CUDA 图形 API 捕获和调用，在 GPU 上实现共轭梯度求解器。

conjugateGradientMultiBlockCG

        这个示例使用多块协作组在 GPU 上实现共轭梯度求解器，还使用统一内存。

conjugateGradientMultiDeviceCG

        这个示例使用多设备协作组在多个 GPU 上实现共轭梯度求解器，还使用通过预取和使用提示优化的统一内存。

conjugateGradientPrecond

        这个示例使用 CUBLAS 和 CUSPARSE 库在 GPU 上实现预处理共轭梯度求解器。

conjugateGradientUM

        这个示例使用 CUBLAS 和 CUSPARSE 库在 GPU 上实现共轭梯度求解器，使用统一内存。

cudaNvSci

        这个示例展示了 CUDA-NvSciBuf/NvSciSync 互操作性。两个 CPU 线程将 NvSciBuf 和 NvSciSync 导入 CUDA，以在 ppm 图像上执行两个图像处理算法——第一个线程中的图像旋转和第二个线程中的旋转图像的 rgba 到灰度转换。目前仅在 Ubuntu 18.04 上支持。

cudaNvSciNvMedia

        这个示例展示了通过 NvSciBuf/NvSciSync API 的 CUDA-NvMedia 互操作性。请注意，该示例仅支持从 x86_64 到 aarch64 的交叉编译，不支持 aarch64 本机编译。有关示例的详细工作流程，请检查示例目录中的 cudaNvSciNvMedia_Readme.pdf。

cuDLAErrorReporting

        这个示例展示了如何通过 CUDA 检测 DLA 错误。

cuDLAHybridMode

        这个示例展示了 DLA 的混合模式，其中 DLA 可以使用 CUDA 进行编程。

cuDLAStandaloneMode

        这个示例展示了 DLA 的独立模式，其中 DLA 可以不使用 CUDA 进行编程。

cuSolverDn_LinearSolver

        一个 CUDA 示例，展示了 cuSolverDN 的 LU、QR 和 Cholesky 分解。

cuSolverRf

        一个 CUDA 示例，展示了 cuSolver 的重分解库 - CUSOLVERRF。

cuSolverSp_LinearSolver

        一个 CUDA 示例，展示了 cuSolverSP 的 LU、QR 和 Cholesky 分解。

cuSolverSp_LowlevelCholesky

        一个 CUDA 示例，展示了使用 cuSolverSP 的低级 API 进行 Cholesky 分解。

cuSolverSp_LowlevelQR

        一个 CUDA 示例，展示了使用 cuSolverSP 的低级 API 进行 QR 分解。

FilterBorderControlNPP

        这个示例展示了如何在启用边界控制的情况下使用 NPP 过滤函数的任何边界版本。提到的函数可以用于复制等效的非边界版本的 NPP 函数的结果。它们还可以用于根据输入图像的不同部分启用和禁用边界控制。

freeImageInteropNPP

        一个简单的 CUDA 示例，展示了如何使用 FreeImage 库与 NPP。

histEqualizationNPP

        这个 CUDA 示例展示了如何使用 NPP 对图像数据进行直方图均衡化。

lineOfSight

        这个示例实现了一个简单的视线算法：给定一个高度图和从某个观察点发出的光线，它计算从观察点可以看到的所有点。实现基于 Thrust 库。

matrixMulCUBLAS

        这个示例实现了编程指南第 3 章中的矩阵乘法。为了展示矩阵乘法的 GPU 性能，该示例还展示了如何使用新的 CUDA 4.0 接口 CUBLAS 实现高性能矩阵乘法。

MersenneTwisterGP11213

        这个示例展示了 cuRAND 中的 Mersenne Twister 随机数生成器 GP11213。

nvJPEG

        一个 CUDA 示例，展示了使用 NVJPEG 库进行单个和批量解码 jpeg 图像。

nvJPEG_encoder

        一个 CUDA 示例，展示了使用 NVJPEG 库对 jpeg 图像进行单个编码。

oceanFFT

        这个示例使用 CUFFT 库模拟海洋高度场，并使用 OpenGL 渲染结果。

randomFog

        这个示例展示了 CURAND 生成的伪随机数和准随机数。

simpleCUBLAS

        使用 CUBLAS API 接口执行 GEMM 操作的示例。

simpleCUBLAS_LU

        CUDA 示例，展示了 cuBLAS API cublasDgetrfBatched() 用于矩阵的 LU 分解。

simpleCUBLASXT

        使用 CUBLAS-XT 库在多个 GPU 上执行 GEMM 操作的示例。

simpleCUFFT

        使用 CUFFT 的示例。在此示例中，CUFFT 用于计算信号与滤波器的 1D 卷积，通过将它们转换到频域，相乘，然后将信号转换回时域。CUFFT 计划使用简单和高级 API 函数创建。

simpleCUFFT_2d_MGPU

        使用 CUFFT 的示例。在此示例中，CUFFT 用于计算信号与滤波器的 2D 卷积，通过将它们转换到频域，相乘，然后将信号转换回时域，多个 GPU 上。

simpleCUFFT_callback

        使用 CUFFT 的示例。在此示例中，CUFFT 用于计算信号与滤波器的 1D 卷积，通过将它们转换到频域，相乘，然后将信号转换回时域。与 Simple CUFFT 示例的区别在于，乘法步骤由 CUFFT 内核使用用户提供的 CUFFT 回调例程完成，而不是通过单独的内核调用完成。

simpleCUFFT_MGPU

        使用 CUFFT 的示例。在此示例中，CUFFT 用于计算信号与滤波器的 1D 卷积，通过将它们转换到频域，相乘，然后将信号转换回时域，多个 GPU 上。

watershedSegmentationNPP

        一个 NPP CUDA 示例，展示了如何使用 NPP 的分水岭分割函数。

5. Domain Specific

        这些示例展示了 CUDA 在图像处理、金融模拟、物理仿真等领域的应用，帮助用户了解如何在特定应用场景中利用 CUDA 技术提高性能和效率。

        特定领域。此部分的示例是针对特定领域的应用，比如图形学、金融、图像处理等。通过这些示例，用户可以了解如何在具体的应用场景中利用 CUDA 技术提高性能和效率。

bicubicTexture

        此示例展示了如何使用 CUDA 纹理高效地实现双三次 B 样条插值滤波器。

bilateralFilter

        双边滤波器是一种保边非线性平滑滤波器，使用 CUDA 实现并通过 OpenGL 渲染。它可以用于图像恢复和去噪。每个像素的加权考虑了其邻居之间的空间距离和颜色距离。参考文献："C. Tomasi, R. Manduchi, Bilateral Filtering for Gray and Color Images, proceeding of the ICCV, 1998, Bilateral Filtering"

binomialOptions

        这个示例评估了一组二项模型下的欧洲期权的公平看涨价格。

binomialOptions_nvrtc

        这个示例评估了一组二项模型下的欧洲期权的公平看涨价格。该示例使用 NVRTC 进行运行时编译。

BlackScholes

        这个示例使用 Black-Scholes 公式评估了一组欧洲期权的公平看涨和看跌价格。

BlackScholes_nvrtc

        这个示例使用 Black-Scholes 公式评估了一组欧洲期权的公平看涨和看跌价格，并使用 NVRTC 进行运行时编译 CUDA 内核。

convolutionFFT2D

        这个示例展示了如何使用 FFT 变换高效地实现具有非常大核尺寸的 2D 卷积。

dwtHaar1D

        对长度为 2 的幂的 1D 信号进行离散 Haar 小波分解。

dxtc

        使用 CUDA 进行高质量的 DXT 压缩。此示例展示了如何在 GPU 上并行实现现有的计算密集型 CPU 压缩算法，并获得数量级的性能提升。

fastWalshTransform

        自然（Hadamard）排序的快速 Walsh 变换，用于处理大小为 2 的幂的任意合适长度的矢量批处理。

FDTD3d

        这个示例在 3D 表面上应用有限差分时域进展模板。

fluidsD3D9

        使用 CUDA 和 CUFFT 进行流体模拟的示例，并通过 Direct3D 9 渲染。需要 Direct3D 兼容设备。

fluidsGL

        使用 CUDA 和 CUFFT 进行流体模拟的示例，并通过 OpenGL 渲染。

fluidsGLES

        使用 CUDA 和 CUFFT 进行流体模拟的示例，并通过 OpenGLES 渲染。

HSOpticalFlow

        变分光流估计算法示例。使用纹理进行图像操作。展示了如何使用 CUDA 加速简单的 PDE 求解器。

Mandelbrot

        此示例使用 CUDA 交互计算并显示 Mandelbrot 或 Julia 集。它还展示了如何使用"double single" 算术在图案中进行长距离缩放时提高精度。此示例使用双精度。感谢 NewTek 的 Mark Granger 提供此代码示例。

marchingCubes

        这个示例使用等值面提取算法从体积数据集中提取几何等值面。它使用 Thrust 库中的扫描（前缀和）函数进行流压缩。

MonteCarloMultiGPU

        这个示例使用蒙特卡罗方法评估一组欧洲期权的公平看涨价格，利用系统中所有支持 CUDA 的 GPU。此示例在存在 GTX 200 类 GPU 时使用双精度硬件。该示例还利用 CUDA 4.0 功能支持使用单个 CPU 线程控制多个 GPU。

nbody

        这个示例展示了高效的全对全重力 n 体模拟。这个示例伴随 GPU Gems 3 章节 "Fast N-Body Simulation with CUDA"。在 CUDA 5.5 中，Tesla K20c 的单精度性能已提高到超过 1.8TFLOP/s。双精度性能在所有 Kepler 和 Fermi GPU 架构上也有所提高。从 CUDA 4.0 开始，nBody 示例已更新为利用新功能在单个 PC 中跨多个 GPU 轻松扩展 n 体模拟。添加“-numbodies=”到命令行将允许用户设置模拟体数。添加“–numdevices=”到命令行选项将使示例使用 N 个设备（如果可用）进行模拟。在这种模式下，所有体的位置信息和速度数据从系统内存中读取使用“零复制”而不是从设备内存中读取。对于少量设备（4 个或更少）和足够大的体数，带宽不是瓶颈，因此我们可以在这些设备之间实现强扩展。

nbody_opengles

        这个示例展示了高效的全对全重力 n 体模拟。与 OpenGL nbody 示例不同，没有用户交互。

nbody_screen

        这个示例展示了高效的全对全重力 n 体模拟。与 OpenGL nbody 示例不同，没有用户交互。

NV12toBGRandResize

        该代码展示了两种使用 CUDA 将 NV12 帧转换并调整大小为 BGR 三平面帧的方法。方法 1，将 NV12 输入转换为 BGR @ 输入分辨率 1，然后调整大小到分辨率 2。方法 2，将 NV12 输入调整大小到分辨率 2，然后将其转换为 BGR 输出。NVIDIA HW 解码器，dGPU 和 Tegra 通常输出 NV12 pitch 格式帧。对于使用 TensorRT 的推理，输入帧需要是 BGR 平面格式，并且可能具有不同的大小。因此，通常需要在解码后进行 NV12 到 BGR 平面的转换和调整大小。此 CUDA 代码提供了转换和调整大小的参考实现。

p2pBandwidthLatencyTest

        这个应用程序展示了 GPU 对之间的 CUDA 点对点 (P2P) 数据传输，并计算延迟和带宽。测试在使用 P2P 和不使用 P2P 的 GPU 对上进行。

postProcessGL

        这个示例展示了如何使用 CUDA 对 OpenGL 渲染的图像进行后处理。

quasirandomGenerator

        这个示例实现了 Niederreiter 准随机序列生成器和逆累积分布函数，用于生成标准正态分布。

quasirandomGenerator_nvrtc

        这个示例实现了 Niederreiter 准随机序列生成器和逆累积分布函数，用于生成标准正态分布，并使用 NVRTC 进行运行时编译 CUDA 内核。

recursiveGaussian

        这个示例使用 Deriche 的递归方法实现了高斯模糊。此方法的优点在于执行时间与滤波器宽度无关。

simpleD3D10

        展示了 CUDA 与 Direct3D10 之间互操作性的简单程序。程序使用 CUDA 生成顶点数组，并使用 Direct3D10 渲染几何图形。需要 Direct3D 兼容设备。

simpleD3D10RenderTarget

        展示了 Direct3D10 渲染目标与 CUDA 的互操作性简单程序。程序使用 CUDA 处理渲染目标位置并生成直方图进行可视化。需要 Direct3D10 兼容设备。

simpleD3D10Texture

        展示了如何与 Direct3D10 纹理进行互操作的简单程序。程序创建了一些由 CUDA 内核生成的 D3D10 纹理（2D、3D 和立方图）。然后，Direct3D 在屏幕上渲染结果。需要 Direct3D10 兼容设备。

simpleD3D11

        展示了如何使用 CUDA D3D11 外部资源互操作性 API 更新 D3D11 缓冲区并使用 Keyed Mutexes 在 D3D11 和 CUDA 之间同步的简单程序。

simpleD3D11Texture

        展示了 Direct3D11 纹理与 CUDA 互操作性的简单程序。程序创建了一些由 CUDA 内核写入的 D3D11 纹理（2D、3D 和立方图）。然后，Direct3D 在屏幕上渲染结果。需要 Direct3D 兼容设备。

simpleD3D12

        一个展示 Direct3D12 与 CUDA 互操作性的程序。程序在 CUDA 内核中创建 DX12 顶点缓冲区中的正弦波，并使用 DirectX12 栅栏在 DX12 和 CUDA 之间进行同步。然后，Direct3D 在屏幕上渲染结果。需要 Windows 10 或更高操作系统上的 DirectX12 兼容 NVIDIA GPU。

simpleD3D9

        展示了 CUDA 与 Direct3D9 之间互操作性的简单程序。程序使用 CUDA 生成顶点数组，并使用 Direct3D9 渲染几何图形。需要 Direct3D 兼容设备。

simpleD3D9Texture

        展示了 Direct3D9 纹理与 CUDA 互操作性的简单程序。程序创建了一些由 CUDA 内核写入的 D3D9 纹理（2D、3D 和立方图）。然后，Direct3D 在屏幕上渲染结果。需要 Direct3D 兼容设备。

simpleGL

        展示了 CUDA 与 OpenGL 之间互操作性的简单程序。程序使用 CUDA 修改顶点位置，并使用 OpenGL 渲染几何图形。

simpleGLES

        展示了 CUDA 与 OpenGL ES 之间的数据交换（即图形互操作）。程序使用 CUDA 修改顶点位置，并使用 OpenGL ES 渲染几何图形。

simpleGLES_EGLOutput

        展示了 CUDA 与 OpenGL ES 之间的数据交换（即图形互操作）。程序使用 CUDA 修改顶点位置，并使用 OpenGL ES 渲染几何图形，并展示了如何使用 EGLOutput 机制和 DRM 库直接渲染到显示器。

simpleGLES_screen

        展示了 CUDA 与 OpenGL ES 之间的数据交换（即图形互操作）。程序使用 CUDA 修改顶点位置，并使用 OpenGL ES 渲染几何图形。

simpleVulkan

        这个示例展示了 Vulkan CUDA 互操作性。CUDA 导入 Vulkan 顶点缓冲区并对其进行操作以创建正弦波，并通过 Vulkan 信号量导入 CUDA 进行同步。此示例依赖于 Vulkan SDK、GLFW3 库，构建此示例请参阅示例目录中的“Build_instructions.txt”。

simpleVulkanMMAP

        这个示例展示了通过 cuMemMap API 进行的 Vulkan CUDA 互操作性。CUDA 导出缓冲区，Vulkan 导入为顶点缓冲区。CUDA 调用内核对顶点进行操作，并通过 Vulkan 信号量导入 CUDA 进行同步。此示例依赖于 Vulkan SDK、GLFW3 库，构建此示例请参阅示例目录中的“Build_instructions.txt”。

SLID3D10Texture

        展示了使用 SLI 与 CUDA 互操作的 Direct3D10 纹理的简单程序。程序创建了一个由 CUDA 内核写入的 D3D10 纹理。然后，Direct3D 在屏幕上渲染结果。需要 Direct3D 兼容设备。

smokeParticles

        使用半角切片技术进行体积阴影烟雾模拟。使用 CUDA 进行程序模拟，使用 Thrust 库进行排序算法，并使用 OpenGL 进行图形渲染。

SobelFilter

        这个示例实现了 8 位单色图像的 Sobel 边缘检测滤波器。

SobolQRNG

        这个示例实现了 Sobol 准随机序列生成器。

stereoDisparity

        一个 CUDA 程序，展示了如何使用 SIMD SAD（绝对差之和）内在函数计算立体视差图。需要计算能力 2.0 或更高。

VFlockingD3D10

        这个示例模拟了由大鸟（如鹅和鹤）形成的 V 形鸟群。该鸟群算法借鉴了 "V-like formations in flocks of artificial birds" 论文中的算法（出自 Artificial Life, Vol. 14, No. 2, 2008）。该示例具有 CPU 和 GPU 实现。按 'g' 键在它们之间切换。基于 GPU 的模拟比基于 CPU 的模拟快很多倍。控制台窗口中的打印输出报告每步模拟时间。按 'r' 键重置鸟的初始分布。

volumeFiltering

        这个示例展示了使用 3D 纹理和 3D 表面写入进行 3D 体积滤波。

volumeRender

        这个示例展示了使用 3D 纹理的基本体积渲染。

vulkanImageCUDA

        这个示例展示了 Vulkan 图像 - CUDA 互操作性。CUDA 导入 Vulkan 图像缓冲区，对其进行方框滤波，并通过 Vulkan 信号量导入 CUDA 进行同步。此示例依赖于 Vulkan SDK、GLFW3 库，构建此示例请参阅示例目录中的“Build_instructions.txt”。

6. Performance

        这些示例展示了在 CUDA 编程中如何通过优化内存对齐、选择合适的内存类型和传输方式来提高数据传输和计算的性能，从而实现高效的 GPU 编程。

        性能优化。这一部分的示例展示了如何进行性能优化，包括代码优化、内存优化、计算优化等。通过这些示例，用户可以学习到各种优化技术，以提升 CUDA 程序的运行效率。

alignedTypes

        这个简单的测试展示了对齐和未对齐结构之间巨大的访问速度差异。它测量对齐和未对齐结构在大数据块上每个元素复制的吞吐量。

transpose

        这个示例展示了矩阵转置的实现。通过不同的方法展示了实现高性能的方法。

UnifiedMemoryPerf

        这个示例通过矩阵乘法内核演示了使用和不使用提示的统一内存性能比较，以及其他类型内存（如零复制缓冲区、分页内存、页锁定内存）在单个 GPU 上执行同步和异步传输的性能表现。

7. libNVVM

        这里的示例展示了如何使用 libNVVM 和 NVVM IR（NVIDIA CUDA Compiler Intermediate Representation）。这些工具用于高级的 CUDA 编译和优化，通过理解和使用这些工具，用户可以实现更高级的代码优化和性能调优。

示例介绍

这些示例展示了 libNVVM 和 NVVM IR 的使用。

• cuda-shared-memory：目录包含展示 CUDA 'shared' 内存使用的 NVVM IR 程序。

• cuda-c-linking：使用 LLVM IR 构建 API 构建 NVVM IR 程序。将生成的 PTX 与 nvcc 生成的 PTX 链接，并使用 CUDA 驱动程序 API 在 GPU 上启动链接的程序。

• device-side-launch：展示在内核内启动另一个内核的功能（CUDA 动态并行）。

• ptxgen：独立的 NVVM IR 到 PTX 编译器。将 libDevice 库与输入 NVVM IR 程序链接，验证 IR 是否符合 NVVM IR 规范，然后生成 PTX。

• simple：从文件中读取 NVVM IR 程序，将其编译为 PTX，并使用 CUDA 驱动程序 API 在 GPU 上启动程序。

• syscalls：目录包含展示设备端 malloc/free/vprintf 函数使用的 NVVM IR 程序。

• uvmlite：展示统一虚拟内存的使用。

构建示例的步骤

以下环境变量可用于控制示例的构建过程。如果未指定，将通过在 PATH 中查找 nvcc 来派生 CUDA_HOME。CMake 将尝试自动识别所有这些路径。

• CUDA_HOME: CUDA 工具包安装目录，例如 /usr/local/cuda。
• LIBNVVM_HOME: libNVVM 组件所在的目录，例如 $CUDA_HOME/nvvm。
• LLVM_HOME: 如果在本地构建了 LLVM，则应指向安装目录。仅在构建 cuda-c-linking 示例时需要（见下文的 cuda-c-linking 注释）。

设置环境变量并通过 PATH 环境变量添加 CMake 工具路径后，可以执行示例脚本 utils/build.sh（适用于 Linux）或 utils/build.bat（适用于 Windows）。此脚本将使用构建目录“build”来构建示例，然后将它们安装在“install”目录中。

如果选择使用 Visual Studio 及其集成的 CMake 支持进行构建，则只需运行“Build All”和“Install libnvvm-samples”。安装的示例将被复制到 out/install/<build architecture>/bin/ 目录中。

另外，我们提供一个 Makefile，作为顶级 cuda-samples 构建的一部分，自动在 Linux 上构建这些示例。Windows 用户应通过 utils/build.bat 或 Visual Studio 的 CMake 集成手动构建。

关于 cuda-c-linking 示例的说明

这个示例需要版本在 7 到 14 之间的开发包（或本地构建的）LLVM 库。LLVM 15 默认使用不透明指针，当前 libNVVM 不支持。

对于希望构建 cuda-c-linking 示例并希望使用本地构建的 LLVM 的用户，需要设置 LLVM_HOME 环境变量。该示例需要包含 LLVM 头文件和库的 LLVM 开发包。

如果满足 LLVM 依赖项，用户可以通过在 CMake 命令行调用中设置 CMake 变量“ENABLE_CUDA_C_LINKING_SAMPLE”或修改此目录中的 CMakeLists.txt 来启用此示例的构建。

Windows 用户应从 llvm.org 下载 LLVM 14 源代码并在本地构建和安装 LLVM。使用 llvm.org 提供的 Windows 安装程序缺少 cuda-c-linking 示例所需的某些组件。

对于 Ubuntu 用户，“llvm-dev”包包含此示例所需的 LLVM 头文件和库，在这种情况下用户不需要显式定义 LLVM_HOME。

Windows 用户应使用与构建 LLVM 相同的 CMake 构建模式来构建此示例。例如，如果他们在 Release 模式下构建了 LLVM，则此示例也应在 Release 模式下构建。可以更新 utils/build.bat 以反映这一点：将“-DCMAKE_BUILD_TYPE=Release”添加到 CMake 调用中。