logo资料库

nv-cuda编程手册中文版.docx

发布时间：2022-05-29 发布人：admin 分类：说明书 资料大小：2.56M 资料格式：docx 举报 版权申诉
第1页 / 共49页
第2页 / 共49页
第3页 / 共49页
第4页 / 共49页
第5页 / 共49页
第6页 / 共49页
第7页 / 共49页
第8页 / 共49页
资料共49页，剩余部分请下载后查看
    第1章 CUDA 介绍
    1.1     作为数据并行计算设备的图像处理器单元
    第2章 编程模型2.1     高度多线程协处理器
    2.2.1    线程块
    2.3    内存模型在设备上执行的线程只能通过下列内存空间访问设备的DRAM和芯片内存储单元，如
    第3章  硬件实现
    3.1     具有芯片共享内存的一组SIMD多处理器
    3.2     执行模型
    第4章  应用编程接口
    4.1     C编程语言扩展
         C语言的最小扩展集合，如4.2所述，允许程序员定位要在设备上执行的部分源码； 运行时ru
    4.2     语言扩展
    4.2.1    函数类型限定符
    4.2.1.1   __device__
    4.2.1.2  __global__
    4.2.1.3   __host__
    4.2.1.4 限制
    4.2.2    变量类型限定符
    4.2.2.1   __device__
    4.2.2.2   __constant__
    4.2.2.3   __shared__
     4.2.2.4   限制
    4.2.3    执行配置
    4.2.4    内置变量
    4.2.4.1   gridDim 
    4.2.4.2   blockIdx
    4.2.4.3 blockDim
    4.2.4.4 threadIdx
    4.2.4.5 限制
    4.2.5    使用NVCC编译
    4.2.5.1   __noinline__
    4.2.5.2   #pragma unroll
    4.3     公共运行时组件
    4.3.1    内置向量类型
    4.3.1.1  char1、uchar1、char2、uchar2、char3、uchar3、c
    4.3.1.2   dim3类型
    4.3.2    数学函数
    4.3.3    时间函数
    4.3.4    纹理类型
    4.3.4.1   纹理参考声明
    4.3.4.2   Runtime运行时纹理参考属性
    4.3.4.3   纹理来自线性内存对来自CUDA数组
     4.4    设备Runtime组件
    4.4.1    数学函数
    4.4.2    同步函数
    4.4.3    类型转换函数
    4.4.4    类型转换函数
    4.4.5    纹理函数
    4.4.5.1   从设备内存取纹理
    4.4.5.2   从CUDA数组取纹理
    4.4.6    原子函数(Atomic Functions)
    4.5    主机Runtime组件
    4.5.1    常用概念
    4.5.1.1   设备
    4.5.1.2   内存
    4.5.1.3   OpenGL互操作性
    4.5.1.4  Direct3D互操作性
    4.5.1.5    异步并发执行 
    4.5.2    Runtime API
    4.5.2.1   初始化
    4.5.2.2   设备管理
    4.5.2.3   内存管理
    4.5.2.4   流管理
    4.5.2.5   事件管理
    4.5.2.6  纹理参考管理
    4.5.2.7   OpenGL互操作性
    4.5.2.8   Direct3D互操作性
    4.5.2.9  使用设备仿真模式调试
    4.5.3    驱动程序API
    表4-1. CUDA驱动程序API中的可用对象 
    4.5.3.1    初始化
    4.5.3.2    设备管理
    4.5.3.3   上下文管理
    4.5.3.4  模块管理
    4.5.3.5   执行控制
    4.5.3.6   内存管理
    4.5.3.7   流管理
    4.5.3.8  事件管理
    4.5.3.9  纹理参考管理
    4.5.3.10  OpenGL互操作性
    4.5.3.11   Direct3D互操作性
    第5章  性能指南
    5.1     指令性能
    5.1.1  指令吞吐量
    5.1.1.1   算术指令
    5.1.1.2   控制流指令
    5.1.1.3   内存指令
    5.1.1.4   同步指令
    5.1.2    内存带宽
    5.1.2.1    全局内存
    5.1.2.2   常量内存
    5.1.2.3   纹理内存
    5.1.2.4   共享内存
    5.1.2.5   寄存器
    5.2     每块的线程数
    5.3     主机和设备之间的数据传送
    5.4     纹理拾取与全局或常量内存读取
     第6章  矩阵乘法的示例
    6.1     概述
    6.2     源码清单
    6.3     源码攻略
    6.3.1    Mul()
    6.3.2    Muld()
    第 1 章 CUDA 介绍 1.1 作为数据并行计算设备的图像处理器单元 仅仅几年的时间，可编程图像处理器单元已经发展成为绝对的计算主力，如图 1-1 所示。由于具有由高内存带宽驱动的多个核心， 今天的 GPU 为图像和非图像处理提供了难以置信的资源。 发展背后的主要原因是 GPU 是特定于计算密集的、高并行的计算，而这正是图像渲染所需要的，因此 GPU 设计了更多的晶体管 专用于数据处理，而非数据高速缓存和流控制，如图 1-2 所示。 具体来说，GPU 专用于解决数据并行计算（同一程序在许多数据元素上并行执行）、具有高运算密度（算术运算与内存操作的比 例）的问题。因为同一程序为每个数据元素执行，所以对高级流控制具有较低的要求；又因为在许多数据元素上执行并具有高运 算密度，所以内存访问延迟可以使用计算而非大的数据高速缓存来隐藏。 数据并行处理将数据元素映射到并行处理线程。处理大型数据集合（比如数组）的许多应用程序可以使用数据并行编程模型来加 速计算。在 3D 渲染中，大型像素和顶点集合映射到并行线程。同样的，图像和媒体处理应用程序，比如渲染图像的后期处理、 视频编码和解码、图像扩展、立体视觉、模式识别，可以将图像块和像素映射到并行处理线程。事实上，图像渲染和处理以外的 许多算法也是可由数据并行处理来加速，范围涉及一般的信号处理或物理模拟，直至财务计算或计算生物学。 但是，直到现在，获得 GPU 中的所有计算能力并将其有效用于非图像应用程序中仍然是一个难题： GPU 只能通过图像 API 进行编程，从而把较高的学习曲线强加给新手，并且为非图像应用程序增加了不必要的 API 开销。 GPU DRAM 可以使用一般方式来读取，即 GPU 程序可以从 DRAM 的任何部分收集数据元素，但不能使用一般方式来写入， 即 GPU 程序不能将消息分散到 DRAM 的任何部分，这就大大降低了 CPU 上易于获取的编程灵活性。 一些应用程序受到 DRAM 内存带宽的瓶颈限制，未充分利用 GPU 的计算功能。 本文档描述了一种新颖的硬件和编程模型，直接解决这些问题，并将 GPU 暴露为一种真正通用的数据并行计算设备。
    1.2 CUDA：一种在 GPU 上进行计算的新架构 CUDA 表示 Compute Unified Device Architecture（统一计算设备架构），是一种新型的硬件和软件架构，用于将 GPU 上作为数 据并行计算设备在 GPU 上进行计算的发放和管理，而无需将其映射到图像 API。它可用于 GeForce 8 系列、Tesla 解决方案和一 些 Quadro 解决方案（详细信息请参阅附录 A）。操作系统的多任务机制负责管理多个并发运行的 CUDA 和图像应用程序对 GPU 的访问。 CUDA 软件堆栈由几层组成，如图 1-3 所示：硬件驱动器，应用编程接口(API)及其 runtime 库，还有两个更高层的通用数学库 CUFFT 和 CUBLAS，这两个库在单独的文档中介绍。硬件已经设计为支持轻量级驱动和 runtime 层，以达到高性能。 CUDA API 包括了对 C 编程语言的扩展，以达到最低的学习曲线（参见第 4 章）。 CUDA 提供了一般的 DRAM 内存寻址以实现更多的编程灵活性，如图 1-4 所示：分散和收集内存操作。从编程角度看，这转换为 就像在 CPU 上一样在 DRAM 的任何位置读取和写入数据的能力。
    CUDA 提供了具有非常快速的一般读写访问的并行数据高速缓存或芯片共享内存，线程可以使用它来互相共享数据（参见第 3 章）。 如图 1-5 所示，应用程序可以利用它来最小化对 DRAM 的过度提取和巡回，从而降低对 DRAM 内存带宽的依赖程度。
    第 2 章 编程模型 2.1 高度多线程协处理器 通过 CUDA 编程时，将 GPU 看作可以并行执行非常多个线程的计算设备(compute device)。它作为主 CPU 的协处理器或者主机 （host）来运作：换句话说，在主机上运行的应用程序中数据并行的、计算密集的部分卸载到此设备上。 更准确地说，多次但在不同数据上独立执行的应用程序部分可以独立放到在此设备上作为许多不同线程执行的函数中。要达到这 种效果，可以将这样一个函数编译到设备的指令集合中，并将得到的程序（叫做内核, kernel）下载到设备上。 主机和设备都保留自己的 DRAM，分别称为主机内存(host memory)和设备内存(device memory)。用户可以通过优化的 API 调用 将数据从一个 DRAM 复制到其他 DRAM 中，而优化的 API 调用使用了设备的高性能直接内存访问(DMA)引擎。 2.2 线程分批 执行内核的线程批组织为线程块的网格，如 2.2.1 和 2.2.2 所述，并参见图 2-1。 2.2.1 线程块 线程块是可以一起协作的线程批次，它们通过一些快速的共享内存有效地共享数据，并同步其执行以协调内存访问。更准确地说， 用户可以在内核中指定同步点，块中的线程在到达此同步点时挂起。 每个线程由线程 ID(thread ID)标识，这是块中的线程号。为了帮助基于线程 ID 的复杂寻址，应用程序还可以将块指定为任意大 小的二维或三维度组，并使用 2 个或 3 个组件索引来标识每个线程。对于大小(Dx,Dy)为的二维块，索引为(x,y)的线程的线程 ID 为(x+yDx)，对于大小为(Dx,Dy,Dz)的三维块，索引为(x,y,z)的线程的线程 ID 为(x+yDx+zDxDy)。 2.2.2 线程块网格 块可以包含的最大线程数是有限制的。但是，执行相同内核的具有相同维度和大小的块可以分批组合到块网格中，以便可以在单 个内核调用中启动的线程总数变得更大。这是以线程协作的降低为代价的，因为同一网格中不同线程块中的线程不能互相通信和 同步。此模型允许内核有效运行，而不必在具有不同并行能力的各种设备上重新编译：如果设备具有非常少的并行能力，则可以 顺序运行网格的所有块，如果具有很多并行能力，则可以并行运行网格的所有块，通常是二者组合使用。 每个块由其块 ID 标识，这是网格中的块号。为了帮助基于块 ID 的复杂寻址，应用程序还可以将网格指定为任意大小的二维度组， 并使用 2 个组件索引来标识每个块。对于大小(Dx,Dy)为的二维块，索引为(x,y)的块的块 ID 为(x+yDx)。 主机执行一连串对设备的内核调 用。每个内核作为组织为线程块 网格的一批线程来执行。 图 2-1. 线程分批
    2.3 内存模型 在设备上执行的线程只能通过下列内存空间访问设备的 DRAM 和芯片内存储单元，如图 2-2 所示： 1、读写每线程寄存器， 2、读写每线程本地内存， 3、读写每块共享内存， 4、读写每网格全局内存， 5、只读每网格常量内存， 6、只读每网格纹理内存。 全局、常量和纹理内存空间可以通过主机读或写，并永久存在于相同应用程序的内核启动中。 全局、常量和纹理内存空间为不同的内存使用进行了优化（参见 5.1.2.1、5.1.2.2 和 5.1.2.3）。纹理内存还为一些特定的数据格式 提供不同的寻址模式以及数据筛选（参见 4.3.4）。 线程可以通过不同范围的一组内存空间来访问设备的 DRAM 和芯片内存。 图 2-2. 内存模型
    第 3 章 硬件实现 3.1 具有芯片共享内存的一组 SIMD 多处理器 设备作为一组多处理器(multiprocessors)来实现，如图 3-1 所示。每个多处理器具有单指令多数据(SIMD)架构：在任何给定的时 钟周期，多处理器的每个处理器执行相同的指令，但操作在不同的数据上。 每个多处理器具有下列四种类型的芯片内存储器： 1、每个处理器有一组本地 32 位寄存器 2、并行数据高速缓存或称为共享内存(shared memory)，由所有处理器共享并实现共享内存空间 3、只读常量高速缓存(constant cache)，由所有处理器共享并加速从常量内存空间的读取，实现为设备内存的只读区域 4、只读纹理高速缓存(texture cache)，由所有处理器共享并加速从纹理内存空间的读取，实现为设备内存的只读区域 本地和全局内存空间为设备内存的读写区域，且无高速缓存。 每个多处理器通过纹理单位(texture unit)访问纹理高速缓存，其中纹理单位实现 2.3 一节提到的各种寻址模式和数据筛选。 具有芯片共享内存的一组 SIMD 多处理器。 图 3-1. 硬件模型
    3.2 执行模型 线程块网格是通过调度块在多处理器上执行来在设备上执行的。每个多处理器一批接一批地处理块批。一个块仅由一个多 处理器处理，所以共享内存空间驻留在芯片共享内存中，从而导致非常快的内存访问。 因为多处理器的寄存器和共享内存划分给块批的所有线程，所以每个多处理器一批可以处理多少块取决于给定内核每秒需 要多少寄存器以及每块需要多少共享内存。如果每个多处理器没有足够的可用寄存器或共享内存来处理至少一个块，则内 核将无法启动。 在一个批次内并被一个多处理器处理的块称为活动(active)块。每个活动块划分到称为 warp 的 SIMD 线程组中：其中每个 warp 包含相同数量的线程，称为 warp 大小，并以 SIMD 方式由多处理器执行。活动 warp——比如所有活动块中的所有 warp——是 分时的：线程调度器（thread scheduler）定期从一个 warp 切换到另一个 warp，以便最大化多处理器计算资源的使用。半 warp （half-warp）是一个 warp 的第一半或第二半。 块划分为 warp 的方式始终相同；每个 warp 包含线程 ID 连续递增的线程，其中第一个 warp 包含线程 0。2.2.1 一节介绍线程 ID 与块中的线程索引如何相关联。 块中 warp 的执行顺序没有定义，但其执行可以同步以协调全局或共享内存访问，如 2.2.1 所述。 线程块网格中块的执行顺序没有定义，且块之间没有同步机制，所以在网格执行期间，来自同一网格的两个不同块中的线程无法 通过全局内存安全地互相通信。 如果对于 warp 的多个线程，由 warp 执行的非完整（non-atomic）指令写入全局或共享内存中的相同位置，则此位置发生的序列 化写入数目及其发生顺序没有定义，但会保证其中一个写入成功。如果由 warp 执行的完整（atomic）指令（参见 4.4.6）读取、 修改并写入 warp 多个线程的全局内存中的相同位置，则对此位置的每个读取、修改和写入都会发生，且全部序列化，但发生的 顺序没有定义。 3.3 计算能力 设备的计算能力(compute capability)由主要修订号和次要修订号来定义。 具有相同主要修订号的设备具有相同的核心架构。附录 A 中列出的设备都具有计算能力 1.x（其主要修订号为 1）。 次要修订号与核心结构的增量改进相对应，其中可能包括新功能。 各种计算能力的技术规范在附录 A 中给出。 3.4 多个设备 使用多个 GPU 作为 CUDA 设备在多 GPU 系统上由应用程序运行时，仅当这些 GPU 具有相同的类型时，才能保证工作。但是， 如果系统处于 SLI 模式，则只有一个 GPU 可以用作 CUDA 设备，因为在驱动器堆栈的最低层，所有 GPU 都熔合在一起。需要 在控制面板中关闭 SLI 模式，CUDA 才能将每个 GPU 看作单独的设备。 3.5 模式开关 GPU 将一些 DRAM 内存指定为所谓的主表面（primary surface），主表面用于用户查看的显示设备的显示刷新。当用户通过更 改显示的分辨率或位深度（使用 NVIDIA 控制面板或 Windows 中的显示控制面板）来启动显示的模式开关（mode switch）时， 主表面所需的内存量将随之变化。例如，如果用户将显示分辨率从 1280x1024x32 位更改为 1280x1024x32 位时，系统必须将 7.68MB 而非 5.24MB 指定为主表面。（运行时启用了反混淆的全屏图像应用程序需要更多的显示内存作为主表面。）在 Windows 中，可以启动显示模式开关的其他事件包括启动全屏 DirectX 应用程序、按下 Alt+Tab 将任务切换出全屏 DirectX 应用程序或按 下 Ctrl+Alt+Del 锁定计算机。 如果模式开关增加主要表明所需的内存量，系统可能必须抽调指定给 CUDA 应用程序的内存分配，从而导致这些应用程序的崩溃。
    第 4 章 应用编程接口 4.1 C 编程语言扩展 CUDA 编程接口的目标是为熟悉 C 编程语言的用户提供相对简单的路径，以便容易地编写在设备上执行的程序。 它包括： C 语言的最小扩展集合，如 4.2 所述，允许程序员定位要在设备上执行的部分源码； 运行时 runtime 库划分为： 主机组件，如 4.5 所述，在主机上运行，提供函数以控制并访问主机中的一个或多个计算设备；  设备组件，如 4.4 所述，在设备上运行，并提供特定于设备的函数；  通用组件，如 4.3 所述，提供内置的向量类型，以及主机和设备代码中都支持的 C 标准库子集。  必须强调一下，只有支持在设备上运行的 C 标准库中的函数才是公共运行时 runtime 组件提供的函数。 4.2 语言扩展 C 编程语言的扩展有四个部分： 函数类型限定符，用于指定函数是在主机上还是在设备上执行，以及可以从主机中还是设备中调用（参见 4.2.1）； 变量类型限定符，用于指定变量在设备上的内存位置（参见 4.2.2）； 新指令，用于指定如何从主机中的设备上执行内核（参见 4.2.3）； 四个内置变量，用于指定网格和块维度，以及块和线程索引（参见 4.2.4）。 包含这些扩展的每个源文件必须使用 CUDA 编译器 nvcc 编译，4.2.5 中有简单介绍。nvcc 的详细介绍可以参见单独的文档。 其中每个扩展附带下文各节中描述的一些限制。违反这些限制时，nvcc 将给出错误或警告，但其中一些违规无法发现。 4.2.1 函数类型限定符 4.2.1.1 __device__ __device__限定符声明函数： 在设备上执行， 只能从设备中调用。 4.2.1.2 __global__
    分享到：
    赞 收藏

    相关推荐

    资料库

    开发技术

    共收录19306份资料，累计14个分类，关注成员有19位，主要包括：Actionscript,Javascript,PHP,VB,Perl,Delphi,Python,C#,C,Java,C++,Web开发,硬件开发,其它

    热门标签

    Actionscript Javascript PHP VB Perl Delphi Python C# C Java C++ Web开发 硬件开发 其它

    最新资料