CUDA C++編程手冊（總論）

CUDA C++編程手冊（總論）

CUDA C++ Programming Guide

The programming guide to the CUDA model and interface.

Changes from Version 10.0

• Use CUDA C++ instead of CUDA C to clarify that CUDA C++ is a C++ language extension not a C language.
• General wording improvements throughput the guide.
• Fixed minor typos in code examples.
• Updated From Graphics Processing to General Purpose Parallel Computing.
• Added reference to NVRTC in Just-in-Time Compilation.
• Clarified linear memory address space size in Device Memory.
• Clarified usage of CUDA_API_PER_THREAD_DEFAULT_STREAM when compiling with nvccin Default Stream.
• Updated Host Functions (Callbacks) to use cudaLaunchHostFunc instead of the deprecated cudaStreamAddCallback.
• Clarified that 8 GPU peer limit only applies to non-NVSwitch enabled systems in Peer-to-Peer Memory Access.
• Added section IOMMU on Linux.
• Added reference to CUDA Compatibility in Versioning and Compatibility.
• Extended list of types supported by __ldg() function in Read-Only Data Cache Load Function.
• Documented support for unsigned short int with atomicCAS().
• Added section Address Space Predicate Functions.
• Added removal notice for deprecated warp vote functions on devices with compute capability 7.x or higher in Warp Vote Functions.
• Added documentation for __nv_aligned_device_malloc() in Dynamic Global Memory Allocation and Operations.
• Added documentation of cudaLimitStackSize in CUDA Dynamic Parallelism Configuration Options.
• Added synchronization performance guideline to CUDA Dynamic Parallelism Synchronization.
• Documented performance improvement of roundf(), round() and updated Maximum ULP Error Table for Mathematical Standard Functions.
• Updated Performance Guidelines Multiprocessor Level for devices of compute capability 7.x.
• Clarified Shared Memory carve out description in Compute Capability 7.x Shared Memory.
• Added missing Stream CUStream Mapping to Driver API
• Added remark about driver and runtime API inter-operability, highlighting cuDevicePrimaryCtxRetain() in Driver API Context.
• Updated default value of CUDA_​CACHE_​MAXSIZE and removed no longer supported environment variables from CUDA Environment Variables
• Added new Unified Memory sections: System Allocator, Hardware Coherency, Access Counters
• Added section External Resource Interoperability .

CUDA C++程序設計指南             

CUDA模型和接口的編程指南。             

版本10.0的更改              

使用CUDA C++代替CUDA C來說明CUDA C++是一種C++語言擴展而不是C語言。             

一般措詞提高吞吐量指南。             

修復了代碼示例中的小錯誤。             

從圖形處理更新為通用並行計算。             

在即時編譯中添加了對NVRTC的引用。             

澄清了設備內存中的線性內存地址空間大小。             

闡明了使用nvccin DEFAULT STREAM編譯時每個線程的CUDA_API_DEFAULT_流的用法。             

已更新主機函數（回調）以使用cudaLaunchHostFunc，而不是已棄用的cudaStreamAddCallback。             

闡明了8 GPU對等限制僅適用於對等內存訪問中啟用非NVSwitch的系統。             

在Linux上添加了IOMMU部分。             

在版本控制和兼容性中添加了對CUDA兼容性的引用。             

只讀數據緩存加載函數中的ldg（）函數支持的擴展類型列表。             

使用atomicCAS（）支持無符號短整型。             

添加了節地址空間謂詞函數。             

在warp投票功能中具有計算能力7.x或更高版本的設備上添加了已棄用的warp投票功能的刪除通知。             

添加了有關動態全局內存分配和操作中與設備對齊的malloc（）的文檔。             

在CUDA動態並行配置選項中添加了cudaLimitStackSize的文檔。             

為CUDA動態並行同步增加了同步性能指標。             

記錄了roundf（）、round（）和更新的數學標准函數最大ULP錯誤表的性能改進。             

更新了計算能力為7.x的設備的多處理器級性能指南。             

闡明了計算能力7.x共享內存中的共享內存划分描述。             

已將缺少的流CUStream映射添加到驅動程序API             

添加了有關驅動程序和運行時API互操作性的注釋，突出顯示了驅動程序API上下文中的cuDevicePrimaryCtxRetain（）。             

更新了CUDA_​CACHE_​MAXSIZE的默認值，並從CUDA環境變量中刪除了不再支持的環境變量             

添加了新的統一內存部分：系統分配器、硬件一致性、訪問計數器             

添加了“外部資源互操作性”一節。

一．介紹             

1.1 從圖形處理到通用並行計算             

由於市場對實時、高清晰度三維圖形的需求無法滿足，可編程圖形處理器單元（GPU）已經發展成為一個高度並行、多線程、多核處理器，具有巨大的計算能力和非常高的內存帶寬，如圖1和圖2所示。 

 圖1.  CPU和GPU每秒的浮點操作數

 圖2.  CPU和GPU的內存帶寬

 CPU和GPU之間浮點能力差異的原因是，GPU專門用於高度並行的計算——正是圖形渲染的目的——因此設計的晶體管更多地用於數據處理，而不是數據緩存和流控制，如圖3所示。 

 圖3.  GPU將更多的晶體管用於數據處理

這在概念上適用於高度並行計算，因為GPU可以通過計算隱藏內存訪問延遲，而不是通過大型數據緩存和流控制來避免內存訪問延遲。             

數據並行處理將數據元素映射到並行處理線程。許多處理大型數據集的應用程序可以使用數據並行編程模型來加快計算速度。在三維渲染中，大量像素和頂點被映射到平行線程。類似地，圖像和媒體處理應用程序，例如渲染圖像的后處理、視頻編碼和解碼、圖像縮放、立體視覺和模式識別，可以將圖像塊和像素映射到並行處理線程。事實上，許多圖像繪制和處理領域之外的算法都是通過數據並行處理來加速的，從一般的信號處理或物理模擬到計算金融或計算生物學。             

1.2. CUDA®通用並行計算平台及編程模型             

2006年11月，NVIDIA推出了CUDA？通用並行計算平台和編程模型，它利用NVIDIA GPUs中的並行計算引擎以比CPU更高效的方式解決許多復雜的計算問題。             

CUDA附帶了一個軟件環境，允許開發人員使用C++作為高級編程語言。如圖4所示，支持其他語言、應用程序編程接口或基於指令的方法，如FORTRAN、DirectCompute、OpenACC。

 圖4. GPU計算應用。CUDA設計用於支持各種語言和應用程序編程接口。

1.3. 一種可擴展的編程模型             

多核CPU和多核GPU的出現意味着主流處理器芯片現在是並行系統。挑戰在於開發應用程序軟件，透明地擴展其並行性，以利用不斷增加的處理器內核數量，就像3D圖形應用程序透明地將其並行性擴展到具有大量不同內核的多個核心gpu一樣。             

CUDA並行編程模型旨在克服這一挑戰，同時為熟悉C等標准編程語言的程序員保持較低的學習曲線。             

它的核心是三個關鍵的抽象——線程組的層次結構、共享內存和障礙同步——它們只是作為一組最小的語言擴展暴露給程序員。             

這些抽象提供了細粒度數據並行和線程並行，嵌套在粗粒度數據並行和任務並行中。它們指導程序員將問題划分為粗的子問題，這些子問題可以由線程塊獨立並行地解決，而每個子問題又划分為更細的子問題，這些子問題可以由塊內的所有線程協同並行地解決。              這種分解通過允許線程在解決每個子問題時進行協作來保持語言的表達能力，同時實現自動可伸縮性。實際上，每個線程塊都可以按任意順序、並發或順序調度到GPU內的任何可用多處理器上，這樣編譯后的CUDA程序就可以在任何數量的多處理器上執行，如圖5所示，並且只有運行時系統需要知道物理多處理器計數。             

這種可擴展的編程模型允許GPU體系結構通過簡單地擴展多處理器和內存分區的數量來跨越廣泛的市場范圍：從高性能的狂熱者GeForce GPU和專業的Quadro和Tesla計算產品到各種便宜的主流GeForce GPU（請參見啟用CUDA的GPU列表在所有啟用CUDA的GPU中）。 

 圖5. 自動伸縮性

注意：GPU是圍繞流式多處理器（SMs）陣列構建的（有關更多詳細信息，請參閱硬件實現）。多線程程序被划分成獨立執行的線程塊，這樣多處理器的GPU將比少處理器的GPU在更短的時間內自動執行程序。             

1.4. 文件結構             

本文件分為以下章節： 

• Chapter Introduction is a general introduction to CUDA.
• Chapter Programming Model outlines the CUDA programming model.
• Chapter Programming Interface describes the programming interface.
• Chapter Hardware Implementation describes the hardware implementation.
• Chapter Performance Guidelines gives some guidance on how to achieve maximum performance.
• Appendix CUDA-Enabled GPUs lists all CUDA-enabled devices.
• Appendix C++ Language Extensions is a detailed description of all extensions to the C++ language.
• Appendix Cooperative Groups describes synchronization primitives for various groups of CUDA threads.
• Appendix CUDA Dynamic Parallelism describes how to launch and synchronize one kernel from another.
• Appendix Mathematical Functions lists the mathematical functions supported in CUDA.
• Appendix C++ Language Support lists the C++ features supported in device code.
• Appendix Texture Fetching gives more details on texture fetching
• Appendix Compute Capabilities gives the technical specifications of various devices, as well as more architectural details.
• Appendix Driver API introduces the low-level driver API.
• Appendix CUDA Environment Variables lists all the CUDA environment variables.
• Appendix Unified Memory Programming introduces the Unified Memory programming model.                  

第一章緒論是對CUDA的概述。
第二章編程模型概述了CUDA編程模型。             

第三章編程接口描述了編程接口。             

第四章硬件實現描述了硬件實現。             

第五章性能指南提供了一些關於如何實現最大性能的指導。             

附錄CUDA啟用的GPU列出所有CUDA啟用的設備。             

附錄C++語言擴展是對C++語言的所有擴展的詳細描述。             

附錄協作組描述了各種CUDA線程組的同步原語。             

附錄CUDA動態並行描述了如何從一個內核啟動和同步另一個內核。             

附錄數學函數列出了CUDA支持的數學函數。             

附錄C++語言支持列出了設備代碼中支持的C++特性。             

附錄紋理提取提供了有關紋理提取的更多詳細信息             

附錄計算能力給出了各種設備的技術規范，以及更多的架構細節。              

附錄驅動程序API引入了底層驅動程序API。             

附錄CUDA環境變量列出了所有CUDA環境變量。             

附錄統一內存編程介紹了統一內存編程模型。