cccl
CCCL 2.6.1
| 撰稿人指南 | 開發容器 | 不和諧 | Godbolt | Github項目 | 文件 |
|---|
歡迎來到CUDA CORE COMPUTE庫（CCCL），我們的任務是使Cuda更令人愉悅。
該存儲庫將三個必需的CUDA C ++庫統計到一個方便的存儲庫中：
CCCL的目標是為CUDA C ++開發人員提供構建塊，使編寫安全有效的代碼變得更加容易。將這些圖書館融合在一起，簡化您的開發過程，並擴大您利用CUDA C ++力量的能力。有關決定統一這些項目的決定的更多信息，請參見此處的公告。
CUDA Core Compute庫（CCCL）的概念從多年來獨立開發的推力，CUB和LIBCUDACXX項目中有機地增長，具有類似的目標：為CUDA開發人員提供高質量，高性能和易於使用的C ++摘要。自然，這三個項目之間存在很多重疊，很明顯，通過將它們統一到一個存儲庫中，社區將得到更好的服務。
推力是C ++並行算法庫，它啟發了向C ++標準庫引入並行算法的引入。 Throust的高級接口極大地提高了程序員的生產率，同時通過可配置的後端啟用GPU和多核心CPU之間的性能可移植性，這些後端允許使用多個並行編程框架（例如CUDA，TBB和OpenMP）。
幼崽是一個較低級別的CUDA特定庫，旨在在所有GPU架構上進行光速並行算法。除了整個設備的算法外，它還還提供合作算法，例如寬範圍的減少和寬範圍的掃描，為CUDA內核開發人員提供構建塊，以創建光速，自定義內核。
libcudacxx是CUDA C ++標準庫。它提供了在主機和設備代碼中使用的C ++標準庫的實現。此外，它為CUDA特異性硬件功能提供了抽象，例如同步原語，緩存控制，原子等。
CCCL的主要目的是填補標準C ++庫填充標準C ++的類似角色：為CUDA C ++開發人員提供通用，光速工具，從而使他們專注於解決重要的問題。統一這些項目是實現這一目標的第一步。
這是一個簡單的示例，說明了推力，幼崽和libcudacxx的CCCL功能的使用。
它顯示瞭如何使用推力/CUB/LIBCUDACXX實現簡單的平行還原核。每個線程塊使用cub::BlockReduce計算數組子集的總和。然後，使用libcudacxx的cuda::atomic_ref添加原子添加，將每個塊的總和還原為單個值。
然後，它顯示瞭如何使用推力的reduce算法進行相同的還原並比較結果。
試試它在Godbolt上！
# include < thrust/execution_policy.h >
# include < thrust/device_vector.h >
# include < cub/block/block_reduce.cuh >
# include < cuda/atomic >
# include < cuda/cmath >
# include < cuda/std/span >
# include < cstdio >
template < int block_size>
__global__ void reduce (cuda::std::span< int const > data, cuda::std::span< int > result) {
using BlockReduce = cub::BlockReduce< int , block_size>;
__shared__ typename BlockReduce::TempStorage temp_storage;
int const index = threadIdx. x + blockIdx. x * blockDim. x ;
int sum = 0 ;
if ( index < data. size ()) {
sum += data[ index ];
}
sum = BlockReduce (temp_storage). Sum (sum);
if (threadIdx. x == 0 ) {
cuda::atomic_ref< int , cuda::thread_scope_device> atomic_result (result. front ());
atomic_result. fetch_add (sum, cuda::memory_order_relaxed);
}
}
int main () {
// Allocate and initialize input data
int const N = 1000 ;
thrust::device_vector< int > data (N);
thrust::fill (data. begin (), data. end (), 1 );
// Allocate output data
thrust::device_vector< int > kernel_result ( 1 );
// Compute the sum reduction of `data` using a custom kernel
constexpr int block_size = 256 ;
int const num_blocks = cuda::ceil_div (N, block_size);
reduce<block_size><<<num_blocks, block_size>>>(cuda::std::span< int const >( thrust::raw_pointer_cast (data. data ()), data. size ()),
cuda::std::span< int >( thrust::raw_pointer_cast (kernel_result. data ()), 1 ));
auto const err = cudaDeviceSynchronize ();
if (err != cudaSuccess) {
std::cout << " Error: " << cudaGetErrorString (err) << std::endl;
return - 1 ;
}
int const custom_result = kernel_result[ 0 ];
// Compute the same sum reduction using Thrust
int const thrust_result = thrust::reduce (thrust::device, data. begin (), data. end (), 0 );
// Ensure the two solutions are identical
std::printf ( " Custom kernel sum: %d n " , custom_result);
std::printf ( " Thrust reduce sum: %d n " , thrust_result);
assert (kernel_result[ 0 ] == thrust_result);
return 0 ;
}CCCL中的所有內容都是僅標頭的。因此，用戶只需要關注他們如何獲取標頭文件以及如何將其整合到構建系統中。
使用CCCL開始使用CCCL的最簡單方法是通過包括CCCL標頭的CUDA工具包。使用nvcc編譯時，它會自動將CCCL標頭添加到包含路徑中，因此您可以簡單地#include代碼中的任何CCCL標頭，而無需其他配置。
如果使用另一個編譯器進行編譯，則需要更新構建系統的搜索路徑，以指向CTK安裝中的CCCL標頭（例如， /usr/local/cuda/include ）。
# include < thrust/device_vector.h >
# include < cub/cub.cuh >
# include < cuda/std/atomic > 鼓勵想要保持在CCCL開發前沿的用戶使用GitHub的CCCL。支持使用較新版本的CCCL具有較舊版本的CUDA工具包，但不反過來。有關CCCL和CUDA工具包之間兼容性的完整信息，請參見我們的平台支持。
CCCL中的所有內容都是僅標頭的，因此克隆並將其包括在一個簡單的項目中與以下內容一樣容易：
git clone https://github.com/NVIDIA/cccl.git
nvcc -Icccl/thrust -Icccl/libcudacxx/include -Icccl/cub main.cu -o main注意使用
-I而不是-isystem，以避免與CUDA工具包的nvcc隱式包含的CCCL標頭髮生衝突。所有CCCL標頭都使用#pragma system_header來確保警告仍然會被沉默，就像使用-isystem一樣，請參閱＃527，有關更多信息。
只能使用install cmake預設來配置僅生成安裝規則的最小構建：
git clone https://github.com/NVIDIA/cccl.git
cd cccl
cmake --preset install -DCMAKE_INSTALL_PREFIX=/usr/local/
cd build/install
ninja install要在安裝中包括實驗庫，請使用install-unstable預設和構建目錄。
要僅安裝實驗庫，請使用install-unstable-only預設並構建目錄。
CCCL還通過conda-forge頻道提供每個版本的Conda包：
conda config --add channels conda-forge
conda install cccl這將安裝最新的CCCL到Conda環境的$CONDA_PREFIX/include/和$CONDA_PREFIX/lib/cmake/ Directories。 Cmake可以通過find_package(CCCL)發現它，並且可以在Conda環境中使用任何編譯器。有關更多信息，請參閱此簡介Conda-Forge。
如果要使用使用特定CUDA工具包的相同的CCCL版本，例如CUDA 12.4，則可以使用以下方式安裝CCCL
conda config --add channels conda-forge
conda install cuda-cccl cuda-version=12.4 cuda-cccl MetaPackage安裝了使用與cuda-version相對應的CUDA工具包的cccl版本。如果您希望在安裝cuda-cccl後更新到最新的cccl ，請在更新cccl之前卸載cuda-cccl ：
conda uninstall cuda-cccl
conda install -c conda-forge cccl請注意，還有帶有諸如
cuda-cccl_linux-64類的名稱的Conda軟件包。這些軟件包包含CCCL版本作為CUDA工具包的一部分，但旨在由CUDA工具包進行內部使用。改為安裝cccl或cuda-cccl，以與CONDA編譯器兼容。有關更多信息，請參閱CCCL Conda-Forge食譜。
CCCL使用CMAKE用於所有構建和安裝基礎架構，包括測試以及在其他CMAKE項目中鏈接的目標。因此，CMAKE是將CCCL集成到另一個項目的推薦方法。
有關如何使用CMAKE軟件包管理器執行此操作的完整示例，請參閱我們的基本示例項目。
其他構建系統應起作用，但僅測試CMAKE。歡迎簡化將CCCL集成到其他構建系統中的貢獻。
有興趣促進使CCCL更好嗎？查看我們的貢獻指南，以全面概述設置開發環境，進行更改，運行測試並提交PR所需的一切。
目的：本節介紹用戶可以期望CCCL能夠成功地運行的地方。
通常，CCCL應在任何地方都可以使用CUDA工具包，但是，魔鬼在細節上。下面的各節描述了CUDA工具包，主機編譯器和C ++方言的不同版本的支持和測試的詳細信息。
概括：
鼓勵CCCL用戶利用最新的增強功能，並始終使用最新版本的CCCL來“居住”。為了獲得無縫的體驗，您可以獨立於整個CUDA工具包升級CCCL。這是可能的，因為CCCL與當前和以前的主要版本系列中的每個次要CTK版本的最新補丁發行版保持向後兼容性。在某些例外情況下，CUDA工具包版本的最低支持的次要版本可能需要比其主要版本系列中最古老的版本更新。例如，由於CTK 11.0中存在不可避免的編譯器問題，CCCL需要11.x系列中的最低支持版本的11.1版本。
當發布新的主要CTK時，我們放棄了對最古老的支持主要版本的支持。
| CCCL版本 | 支持CUDA工具包 |
|---|---|
| 2.x | 11.1-11.8，12.x（僅最新的補丁發布） |
| 3.x（未來） | 12.x，13.x（僅最新補丁發布） |
使用最新的CCCL的行為良好的代碼應使用任何支持的CTK版本成功編譯並成功運行。取決於新的CTK功能的新功能可能會出現例外，因此這些功能在CTK的較舊版本上不起作用。例如，直到CUDA 12.0之前，C ++ 20的支持才不會添加到nvcc中，因此依賴C ++ 20的CCCL功能將無法與CTK 11.x一起使用。
用戶可以將較新版本的CCCL集成到較舊的CTK中，但反之亦然。這意味著CCCL的較舊版本與較新的CTK不兼容。換句話說， CCCL永遠不會與CUDA工具包兼容。
下表總結了CTK和CCCL的兼容性：
| CTK版本 | 包括CCCL版本 | 所需的CCCL | 支持？ | 筆記 |
|---|---|---|---|---|
CTK XY | CCCL MAJOR.MINOR | CCCL MAJOR.MINOR+n | ✅ | 一些新功能可能不起作用 |
CTK XY | CCCL MAJOR.MINOR | CCCL MAJOR+1.MINOR | ✅ | 可能的休息；一些新功能可能無法使用 |
CTK XY | CCCL MAJOR.MINOR | CCCL MAJOR+2.MINOR | CCCL僅支持兩個CTK主要版本 | |
CTK XY | CCCL MAJOR.MINOR | CCCL MAJOR.MINOR-n | CCCL不兼容 | |
CTK XY | CCCL MAJOR.MINOR | CCCL MAJOR-n.MINOR | CCCL不兼容 |
有關CCCL版本控制，API/ABI兼容性和破壞更改的更多信息，請參見下面的版本控制部分。
除非另有說明，否則CCCL支持與CUDA工具包相同的所有操作系統，這些操作系統已記錄在這裡：
除非另有說明，否則CCCL支持與最新的CUDA工具包相同的主機編譯器，這些編譯器已記錄在此處：
當使用較舊的CUDA工具包時，我們還僅支持最新的CUDA工具包的主機編譯器，但至少是任何受支持的舊CUDA工具包的最新主機編譯器。
我們可能會保留其他編譯器的支持，並將接受通過合理修復的社區的相應補丁。但是，我們不會花費大量時間來為較老的編譯器分類或解決問題。
本著“您只支持您測試的內容”的精神，請參閱我們的CI概述，以獲取有關我們測試的內容的更多信息。
除非另有說明，否則CCCL支持與CUDA工具包相同的所有GPU架構/計算功能，這些功能在此處進行了記錄：https：//docs.nvidia.com/cuda/cuda-cuda-cuda-cuda-cuda-cuda-cuda-cuda-cuda-cuda-cuda-cuda-cuda-cuda-cuda-cuda-cuda-coda-colgramming-guide/index.html#compute-copute-capapute-compute-copute-capapiality
請注意，某些功能只能支持某些架構/計算功能。
CCCL的測試策略在測試盡可能多的配置與保持合理的CI時間之間取得了平衡。
對於CUDA工具包，對最古老和最新支持的版本進行了測試。例如，如果最新版本的CUDA工具包為12.3，則針對11.1和12.3進行測試。對於每個CUDA版本，構建均可完成所有受支持的主機編譯器，並具有所有受支持的C ++方言。
測試策略和矩陣不斷發展。 ci/matrix.yaml中定義的矩陣。YAML文件是真實的確定來源。有關CI管道的更多信息，請參見此處。
目的：本節介紹了CCCL的版本，API/ABI穩定性保證以及兼容指南，以最大程度地減少升級頭痛。
概括
cub::或thrust::名稱空間cuda::名稱空間中符號的變化，可能會隨時發生，但會通過將ABI版本嵌入到所有cuda::符號的內聯命名空間中來反映。可以同時支持多個ABI版本。注意：將推力，CUB和LIBCUDACXX合併到此存儲庫之前，每個庫都是根據語義版本獨立版本的。從2.1版本開始，所有三個庫都在單獨的存儲庫中同步了其發行版本。向前邁進，CCCL將繼續以單個語義版本發布，其中2.2.0是NVIDIA/CCCL存儲庫的第一個版本。
打破變化是更改已發布版本之間明確支持的功能，這將需要用戶進行工作以升級到較新版本。
在極限上，任何變化都有可能將某人闖入某個地方。結果，並非所有可能的源打破變化都被認為是打破了對公共API的破壞，保證撞擊主要語義版本。
以下各節描述了破壞CCCL API和ABI的變化的細節。
CCCL的公共API是有意暴露於提供圖書館實用程序的全部功能。
換句話說，CCCL的公共API不僅超越了功能簽名，還包括（但不限於）：
此外，CCCL的公共API不包括以下任何一個：
_或__前綴為前綴的任何符號detail符號，包括detail::名稱空間或宏detail/目錄或子目錄中包含的任何標頭文件總的來說，目標是避免在公共API中破壞任何東西。僅當這些更改為用戶提供更好的性能，更易於理解的API和/或更一致的API時才進行。
對公共API的任何破壞都將需要碰撞CCCL的主要版本編號。為了與CUDA MINIL版本的兼容性保持一致，API破壞變化和CCCL主要版本顛簸只會與CUDA工具包的新的主要版本發行相吻合。
沒有任何警告的任何事情都可能隨時改變公共API。
所有CCCL組件的公共API共享MAJOR.MINOR.PATCH的統一語義版本。
僅支持最近發布的版本。通常，功能和錯誤修復程序不會退回到先前發布的版本或分支。
查詢該版本的首選方法是使用CCCL_[MAJOR/MINOR/PATCH_]VERSION如下所述。為了向後兼容，可以使用推力/CUB/LIBCUDACXXX版本定義，並且將始終與CCCL_VERSION一致。請注意，推力/幼崽使用MMMmmmpp方案，而cccl和libcudacxx則使用MMMmmmppp 。
| CCCL | libcudacxx | 推力 | 幼獸 | |
|---|---|---|---|---|
| 標題 | <cuda/version> | <cuda/std/version> | <thrust/version.h> | <cub/version.h> |
| 主要版本 | CCCL_MAJOR_VERSION | _LIBCUDACXX_CUDA_API_VERSION_MAJOR | THRUST_MAJOR_VERSION | CUB_MAJOR_VERSION |
| 次要版本 | CCCL_MINOR_VERSION | _LIBCUDACXX_CUDA_API_VERSION_MINOR | THRUST_MINOR_VERSION | CUB_MINOR_VERSION |
| 補丁/simbinor版本 | CCCL_PATCH_VERSION | _LIBCUDACXX_CUDA_API_VERSION_PATCH | THRUST_SUBMINOR_VERSION | CUB_SUBMINOR_VERSION |
| 串聯版本 | CCCL_VERSION (MMMmmmppp) | _LIBCUDACXX_CUDA_API_VERSION (MMMmmmppp) | THRUST_VERSION (MMMmmmpp) | CUB_VERSION (MMMmmmpp) |
應用程序二進制接口（ABI）是一組規則：
圖書館的ABI包括但不限於：
ABI破裂的變化是任何會導致函數或鍵入公共API的ABI的變化。例如，將新數據成員添加到結構是ABI破壞變化，因為它會更改類型的大小。
在CCCL中，關於ABI的保證如下：
thrust::和cub::名稱空間中的符號可能隨時打破ABI而不會警告。thrust::和cub::符號的ABI包括用於編譯的CUDA架構。因此thrust::如果cub::-x cu ）與C ++源（ -x cpp ）cuda::名稱空間中的符號也可能隨時破壞ABI。但是， cuda::符號嵌入了一個ABI版本號，每當發生ABI中斷時，該編號會增加。可以同時支持多個ABI版本，因此用戶可以選擇恢復為先前的ABI版本。有關更多信息，請參見此處。誰應該關心阿比？
通常，CCCL用戶只需要擔心在構建或使用二進製文物（例如共享庫）時，其API直接或間接包含CCCL提供的類型。
例如，考慮是否使用CCCL版本X構建libA.so ，其公共API包含一個函數：
void foo (cuda::std::optional< int >);如果使用CCCL版本Y編制了另一個庫libB.so ，並使用libA.so中的foo ，那麼如果版本X和Y之間存在ABI斷裂，則可能會失敗。與API斷開更改不同，ABI斷裂通常不需要代碼更改，只需要重新編譯所有使用相同的ABI版本即可。
要了解有關ABI的更多信息，為什麼它很重要，看看ABI是什麼，C ++應該怎麼做？
如上所述，並非所有可能的源破壞變化都構成了破壞變化，這將需要增加CCCL的API主要版本編號。
鼓勵用戶遵守以下準則，以最大程度地降低意外造成的干擾風險，具體取決於不屬於公共API的CCCL部分：
thrust:: cub:: ， nv:: ，或cuda::名稱空間，例如，例如， cuda::std::iterator_traitsthrust:: cub:: ， cuda::或nv:: namespaces中獲取任何API的地址。thrust:: ” cub:: ， cuda::或nv::名稱空間中的任何API。_ ， __的前綴或名稱中任何地方的detail ，包括detail:: namespace或acro#include聲明該符號的標頭文件。換句話說，不要依靠其他標題隱含包含的標頭。本節的一部分靈感來自Abseil的兼容指南。
在對公共API，ABI或修改支持的平台和編譯器進行任何破壞更改之前，我們將盡最大努力通知用戶。
違反的折舊將以可以禁用的程序性警告的形式出現。
貶值期將取決於變更的影響，但通常至少持續2個次要版本。
即將推出！
有關CI管道的詳細概述，請參見Ci-overview.md。
與CCCL使Cuda更令人愉悅的使命相關的項目：
您的項目使用CCCL嗎？打開PR將您的項目添加到此列表中！
