CUDA编程指南4.0中文版 译者：风辰 由于小弟的水平所限，此文档可能存在错误，如果你觉得本文档的 某些内容可能是错误，请联系我，谢谢！ 由于这样或者那样的原因，此翻译版将可能会是“绝版”，谢谢大家 的支持。 任何人不得更改此文档内容或设置，更不能用于商业目的，本人不 承担任何责任
CUDA 编程指南 4.0 中文版 译者：风辰 由于小弟的水平所限，此文档可能存在错误，如果你觉得本文档的 某些内容可能是错误，请联系我，谢谢！ 由于这样或者那样的原因，此翻译版将可能会是“绝版”，谢谢大家 的支持。 任何人不得更改此文档内容或设置，更不能用于商业目的，本人不 承担任何责任
目录 第一章导论 1.1从图形处理到通用并行计算 1 1.2 CUDATM:一种通用并行计算架构 3 1.3一种可扩展的编程模型 1.4文档结构 第二章编程模型 7 2.1内核 2.2线程层次 .7 2.3存储器层次 10 2.4异构编程 11 2.5计算能力.… 13 第三章编程接口.… 15 3.1用nvcc编译 15 3.11编译流程.… 15 3.1.1.1离线编译 15 3.1.1.2即时编译 16 3.1.2二进制兼容性 。+ 17 3.1.3PTX兼容性 17 3.1.4应用兼容性 17 3.1.5C/C+兼容性 18 3.1.664位兼容性 18 3.2 CUDA C运行时 18 3.2.1初始化 19 3.2.2设备存储器 19 3.2.3共享存储器 22 3.2.4分页锁定主机存储器 28 3.2.4.1可分享存储器（portable memory) 28 3.2.4.2写结合存储器 28 3.2.4.3被映射存储器 28 3.2.5异步并发执行 29 3.2.5.1主机和设备间异步执行 29 3.2.5.2数据传输和内核执行重叠 30 3.2.5.3并发内核执行… 30
目录 第一章 导论......................................................................................................................... 1 1.1 从图形处理到通用并行计算................................................................................ 1 1.2 CUDATM：一种通用并行计算架构...................................................................... 3 1.3 一种可扩展的编程模型........................................................................................ 3 1.4 文档结构................................................................................................................ 5 第二章 编程模型................................................................................................................. 7 2.1 内核........................................................................................................................ 7 2.2 线程层次................................................................................................................ 7 2.3 存储器层次.......................................................................................................... 10 2.4 异构编程.............................................................................................................. 11 2.5 计算能力.............................................................................................................. 13 第三章 编程接口............................................................................................................... 15 3.1 用 nvcc 编译......................................................................................................... 15 3.1.1 编译流程................................................................................................... 15 3.1.1.1 离线编译........................................................................................ 15 3.1.1.2 即时编译........................................................................................ 16 3.1.2 二进制兼容性........................................................................................... 17 3.1.3 PTX 兼容性................................................................................................ 17 3.1.4 应用兼容性............................................................................................... 17 3.1.5 C/C++兼容性 ............................................................................................. 18 3.1.6 64 位兼容性............................................................................................... 18 3.2 CUDA C 运行时 ................................................................................................... 18 3.2.1 初始化....................................................................................................... 19 3.2.2 设备存储器............................................................................................... 19 3.2.3 共享存储器............................................................................................... 22 3.2.4 分页锁定主机存储器............................................................................... 28 3.2.4.1 可分享存储器（portable memory）.............................................. 28 3.2.4.2 写结合存储器................................................................................ 28 3.2.4.3 被映射存储器................................................................................ 28 3.2.5 异步并发执行........................................................................................... 29 3.2.5.1 主机和设备间异步执行................................................................ 29 3.2.5.2 数据传输和内核执行重叠............................................................ 30 3.2.5.3 并发内核执行................................................................................ 30
3.2.5.4并发数据传输 30 3.2.5.5流 30 3.2.5.6事件 33 3.2.5.7同步调用 34 3.2.6多设备系统 34 3.2.6.1枚举设备 34 3.2.6.2设备选择 34 3.2.6.3流和事件行为 35 3.2.6.4p2p存储器访问 35 3.2.6.5p2p存储器复制 36 3.2.7统一虚拟地址空间 37 3.2.8错误检查 37 3.2.9调用栈 38 3.2.10纹理和表面存储器 38 3.2.10.1纹理存储器 38 3.2.10.2表面存储器(surface) 44 3.2.10.3CUDA数组 46 3.2.10.4读写一致性 47 3.2.11图形学互操作性. 47 3.2.11.1 OpenGL互操作性 47 3.2.11.2 Direct3.D互操作性 50 3.2.11.3SLI(速力)互操作性 58 3.3驱动API 58 3.3.1上下文 61 3.3.2模块 62 3.3.3内核执行 63 3.34设备存储器 65 3.3.5共享存储器 69 3.3.6分页锁定主机存储器 71 3.3.7异步并发执行 71 3.3.7.1流. 72 3.3.7.2事件.… 72 3.3.7.3同步调用 73
4 3.2.5.4 并发数据传输................................................................................ 30 3.2.5.5 流.................................................................................................... 30 3.2.5.6 事件................................................................................................ 33 3.2.5.7 同步调用........................................................................................ 34 3.2.6 多设备系统............................................................................................... 34 3.2.6.1 枚举设备........................................................................................ 34 3.2.6.2 设备选择........................................................................................ 34 3.2.6.3 流和事件行为................................................................................ 35 3.2.6.4 p2p 存储器访问.............................................................................. 35 3.2.6.5 p2p 存储器复制.............................................................................. 36 3.2.7 统一虚拟地址空间................................................................................... 37 3.2.8 错误检查................................................................................................... 37 3.2.9 调用栈....................................................................................................... 38 3.2.10 纹理和表面存储器................................................................................. 38 3.2.10.1 纹理存储器.................................................................................. 38 3.2.10.2 表面存储器(surface).................................................................... 44 3.2.10.3 CUDA 数组.................................................................................. 46 3.2.10.4 读写一致性.................................................................................. 47 3.2.11 图形学互操作性..................................................................................... 47 3.2.11.1 OpenGL 互操作性........................................................................ 47 3.2.11.2 Direct3D 互操作性....................................................................... 50 3.2.11.3 SLI（速力）互操作性................................................................. 58 3.3 驱动 API .............................................................................................................. 58 3.3.1 上下文....................................................................................................... 61 3.3.2 模块........................................................................................................... 62 3.3.3 内核执行................................................................................................... 63 3.3.4 设备存储器............................................................................................... 65 3.3.5 共享存储器............................................................................................... 69 3.3.6 分页锁定主机存储器............................................................................... 71 3.3.7 异步并发执行........................................................................................... 71 3.3.7.1 流.................................................................................................... 72 3.3.7.2 事件................................................................................................ 72 3.3.7.3 同步调用........................................................................................ 73
目录 33.8多设备系统 73 3.3.8.1设备枚举 73 3.3.8.2p2p存储器访问 74 3.3.8.3p2p存储器拷贝 74 3.3.9统一虚拟地址空间 75 3.3.10错误检查 75 3.3.11调用栈 76 3.3.12纹理存储器和表面存储器 76 3.3.12.1纹理存储器 76 3.3.12.2表面存储器 78 3.3.13图形学互操作性… 80 3.3.13.1 OpenGL互操作性 80 3.3.10.2 Direct3.D互操作性 82 3.4运行时API和驱动API的互操作性 91 3.5版本和互操作性 91 3.6计算模式 92 3.7模式切换 92 3.8 Windows上的Tesla计算集群模式 93 第四章硬件实现… 95 4.1SIMT架构. 95 4.2硬件多线程 96 第五章性能指南… 99 5.1总体性能优化策略 99 5.2最大化利用率 99 5.2.1应用层次… 99 5.2.2设备层次 99 5.2.3多处理器层次 100 5.3最大化存储器吞吐量 101 5.3.1主机和设备的数据传输 102 5.3.2设备存储器访问 103 5.3.2.1全局存储器 103 5.3.2.2本地存储器 104 5.3.2.3共享存储器 .105
目录 3.3.8 多设备系统............................................................................................... 73 3.3.8.1 设备枚举........................................................................................ 73 3.3.8.2 p2p 存储器访问.............................................................................. 74 3.3.8.3 p2p 存储器拷贝.............................................................................. 74 3.3.9 统一虚拟地址空间................................................................................... 75 3.3.10 错误检查................................................................................................. 75 3.3.11 调用栈 ..................................................................................................... 76 3.3.12 纹理存储器和表面存储器..................................................................... 76 3.3.12.1 纹理存储器.................................................................................. 76 3.3.12.2 表面存储器.................................................................................. 78 3.3.13 图形学互操作性..................................................................................... 80 3.3.13.1 OpenGL 互操作性........................................................................ 80 3.3.10.2 Direct3D 互操作性 ....................................................................... 82 3.4 运行时 API 和驱动 API 的互操作性 ................................................................. 91 3.5 版本和互操作性.................................................................................................. 91 3.6 计算模式.............................................................................................................. 92 3.7 模式切换.............................................................................................................. 92 3.8 Windows 上的 Tesla 计算集群模式..................................................................... 93 第四章 硬件实现............................................................................................................... 95 4.1 SIMT 架构............................................................................................................ 95 4.2 硬件多线程.......................................................................................................... 96 第五章 性能指南............................................................................................................... 99 5.1 总体性能优化策略.............................................................................................. 99 5.2 最大化利用率...................................................................................................... 99 5.2.1 应用层次................................................................................................... 99 5.2.2 设备层次................................................................................................... 99 5.2.3 多处理器层次......................................................................................... 100 5.3 最大化存储器吞吐量........................................................................................ 101 5.3.1 主机和设备的数据传输......................................................................... 102 5.3.2 设备存储器访问..................................................................................... 103 5.3.2.1 全局存储器.................................................................................. 103 5.3.2.2 本地存储器.................................................................................. 104 5.3.2.3 共享存储器.................................................................................. 105
6 5.3.2.4常量存储器 105 5.3.2.5纹理和表面存储器 106 5.4最大化指令吞吐量 106 5.4.1算术指令 106 5.4.2控制流指令 109 5.4.3同步指令 109 附录A支持CUDA的GPU 113 附录BC语言扩展.… 117 B.1函数类型限定符 .117 B.1.1 device 117 B.1.2 global .117 B.1.3 host 117 B.1.4_noinline和 forceinline 118 B.2变量类型限定符 .118 B.2.1 device 118 B.2.2 constant .118 B.2.3 shared .119 B.2.4 restrict 120 B.3内置变量类型 .121 B.3.1 charl、uchar1、char2、uchar.2、char3、uchar.3、char4、uchar44、short1、 ushort1、short2、ushort.2、short.3、ushort3.、shorta4、ushort44、intl、uintl、int2、 uint2、int3、uint3、int4、uint4、long1、ulongl、long2、ulong2、long3、ulong3、 long4、ulong4、float1、float2、float3、float4、double2 121 B.3.2dim3类型 122 B.4内置变量 122 B.4.1 gridDim........... 122 B.4.2 blockIdx.... 122 B.4.3 blockDim 122 B.4.4 threadIdx 122 B.4.5 warpSize 122 B.5存储器栅栏函数 122 B.6同步函数 124 B.7数学函数.… 125
6 5.3.2.4 常量存储器.................................................................................. 105 5.3.2.5 纹理和表面存储器...................................................................... 106 5.4 最大化指令吞吐量............................................................................................ 106 5.4.1 算术指令................................................................................................. 106 5.4.2 控制流指令............................................................................................. 109 5.4.3 同步指令................................................................................................. 109 附录 A 支持 CUDA 的 GPU ...........................................................................................113 附录 B C 语言扩展 ...........................................................................................................117 B.1 函数类型限定符 ................................................................................................117 B.1.1 __device__................................................................................................117 B.1.2 __global__ ................................................................................................117 B.1.3 __host__....................................................................................................117 B.1.4 _noinline_ 和 __forceinline__................................................................118 B.2 变量类型限定符 ................................................................................................118 B.2.1 __device__................................................................................................118 B.2.2 __constant__.............................................................................................118 B.2.3 __shared__................................................................................................119 B.2.4 __restrict__ .............................................................................................. 120 B.3 内置变量类型 ................................................................................................... 121 B.3.1 char1、uchar1、char2、uchar2、char3、uchar3、char4、uchar4、short1、 ushort1、short2、ushort2、short3、ushort3、short4、ushort4、int1、uint1、int2、 uint2、int3、uint3、int4、uint4、long1、ulong1、long2、ulong2、long3、ulong3、 long4、ulong4、float1、float2、float3、float4、double2 ............................ 121 B.3.2 dim3 类型................................................................................................ 122 B.4 内置变量 ........................................................................................................... 122 B.4.1 gridDim.................................................................................................... 122 B.4.2 blockIdx ................................................................................................... 122 B.4.3 blockDim ................................................................................................. 122 B.4.4 threadIdx.................................................................................................. 122 B.4.5 warpSize .................................................................................................. 122 B.5 存储器栅栏函数 ............................................................................................... 122 B.6 同步函数 ........................................................................................................... 124 B.7 数学函数 ........................................................................................................... 125