cuSPARSELt開發NVIDIA Ampere結構化稀疏性
深度神經網絡在各種領域（例如計算機視覺，語音識別和自然語言處理）中均具有出色的性能。處理這些神經網絡所需的計算能力正在迅速提高，因此有效的模型和計算至關重要。神經網絡剪枝（刪除不必要的模型參數以生成稀疏網絡）是一種在保持準確性的同時降低模型復雜性的有用方法。
為了利用細粒度的網絡剪枝，NVIDIA Ampere GPU架構引入了細粒度的結構稀疏性的概念。在NVIDIA A100 GPU上，結構顯示為2：4模式：每四個元素中至少有兩個必須為零。通過使用新的NVIDIA Sparse Tensor Core跳過零值的計算，這可以將一個矩陣乘法（也稱為GEMM）操作數的數據占用空間和帶寬減少2倍，并使吞吐量翻倍。
cuSPARSELt：用于稀疏矩陣-密集矩陣乘法的高性能CUDA庫
為了簡化NVIDIA Ampere架構稀疏功能的使用，NVIDIA引入了cuSPARSELt ，這是一種高性能CUDA庫，專用于常規矩陣操作，其中至少一個操作數是稀疏矩陣。cuSPARSELt庫可以使用NVIDIA第三代Tensor Core稀疏矩陣乘累加（SpMMA）操作，而無需進行底層編程。該庫還提供用于剪枝和壓縮矩陣的輔助函數。
cuSPARSELt的主要功能包括：
· NVIDIA Sparse Tensor Core支持
· 混合精度支持：
o FP16輸入/輸出，FP32張量核心累積
o BFLOAT16輸入/輸出，FP32張量核心累積
o INT8輸入/輸出，INT32張量核心累積
· Row-major and column-major memory layouts的內存布局
· 矩陣剪枝和壓縮實用程序
· 自動調整功能
NVIDIA Sparse Tensor Core support
Mixed-precision support:
FP16 inputs/output, FP32 Tensor Core accumulation
BFLOAT16 inputs/output, FP32 Tensor Core accumulation
INT8 inputs/output, INT32 Tensor Core accumulation
Row-major and column-major memory layouts
Matrix pruning and compression utilities
Auto-tuning functionality
定制工作流程
cuSPARSELt庫遵循等效方法，并采用與cuBLASLt和cuTENSOR類似的概念。庫編程模型要求以某種方式組織計算，以使相同的設置可以重復用于不同的輸入。
該模型尤其依賴于以下高層階段：
· 問題定義：指定矩陣形狀，數據類型，操作等。
· 用戶偏好和約束：提供算法選擇或限制可行實現（候選）的搜索空間。
· 計劃：收集執行的描述符，并在需要時“找到”最佳實施。
· 執行：執行實際計算。
通用工作流程包括以下步驟：

1. 初始化庫句柄：cusparseLtInit。
2. 指定輸入/輸出矩陣特征：cusparseLtDenseDescriptorInit， cusparseLtStructuredDescriptorInit。
3. 初始化矩陣乘法描述符和它的屬性（例如操作，計算類型等）： cusparseLtMatmulDescriptorInit。
4. 初始化算法選擇描述符：cusparseLtMatmulAlgSelectionInit。
5. 初始化矩陣乘法計劃：cusparseLtMatmulPlanInit。
6. 剪枝A矩陣：cusparseLtSpMMAPrune。如果用戶提供已經滿足2：4結構化稀疏性約束的矩陣，例如由ASP庫生成的權重矩陣，則不需要此步驟。
7. 壓縮剪枝后的矩陣：cusparseLtSpMMACompress。
8. 執行矩陣乘法：cusparseLtMatmul。可以使用不同的輸入多次重復此步驟。
9. 銷毀矩陣乘法計劃和庫句柄：cusparseLtMatmulPlanDestroy，cusparseLtDestroy。
   稀疏的GEMM性能
   與密集矩陣乘法一樣，稀疏矩陣乘法的性能隨GEMM尺寸，布局和數據類型而變化。這是當前軟件與稀疏GEMM相對性能的快照。
   下表顯示了cuSPARSELt和cuBLAS在以下操作中的性能：
   D = alpha * op（A）* op（B）+ beta * C
   在該操作中，A，B，和 D=C分別是尺寸的密集矩陣MXK，KXN，和M×N個。矩陣的布局A和B與?為列主順序（OP是非轉置）和?為行優先順序（OP調換）。
   為了展示使用cuSPARSELt可以針對實際工作負載實現的性能，下表顯示了帶有主要列TN FP16內核的剪枝后的BERT-Large模型（seqlen = 128，BS = 128）使用的一些常見GEMM大小。通常，工作量越大，稀疏性可以提供的幫助越多。
   

表1. BERT-Large模型和不同層的cuSPARSELt性能。
結構化稀疏矩陣-矩陣乘法代碼示例
已經看到了可用的性能，下面是一個示例，該示例使用NVIDIA A100或GA100 GPU中的稀疏Tensor內核在cuSPARSELt庫中執行具有結構稀疏性的矩陣乘法。有關更多信息，請參見NVIDIA / CUDALibrarySamples / tree / master / cuSPARSELt / spmma GitHub存儲庫。
首先，包括cuSPARSELt標頭，設置一些設備指針和數據結構，并初始化cuSPARSELt句柄。
#include <cusparseLt.h> // cusparseLt header // Device pointers and coefficient definitions float alpha = 1.0f; float beta = 0.0f; __half* dA = … __half* dB = … __half* dC = … // cusparseLt data structures and handle initialization cusparseLtHandle_t handle; cusparseLtMatDescriptor_t matA, matB, matC; cusparseLtMatmulDescriptor_t matmul; cusparseLtMatmulAlgSelection_t alg_sel; cusparseLtMatmulPlan_t plan; cudaStream_t stream = nullptr; cusparseLtInit(&handle);
接下來，初始化結構化的稀疏輸入矩陣（matrix A），密集輸入矩陣（matrix B）和密集輸出矩陣（matrix C）描述符。
cusparseLtStructuredDescriptorInit(&handle, &matA, num_A_rows, num_A_cols, lda, alignment, type, order, CUSPARSELT_SPARSITY_50_PERCENT); cusparseLtDenseDescriptorInit(&handle, &matB, num_B_rows, num_B_cols, ldb, alignment, type, order); cusparseLtDenseDescriptorInit(&handle, &matC, num_C_rows, num_C_cols, ldc, alignment, type, order);
準備好描述符后，可以準備矩陣乘法運算的描述符，選擇用于執行matmul運算的算法，并初始化matmul計劃。
cusparseLtMatmulDescriptorInit(&handle, &matmul, opA, opB, &matA, &matB, &matC, &matC, compute_type); cusparseLtMatmulAlgSelectionInit(&handle, &alg_sel, &matmul, CUSPARSELT_MATMUL_ALG_DEFAULT); int alg = 0; // set algorithm ID cusparseLtMatmulAlgSetAttribute(&handle, &alg_sel, CUSPARSELT_MATMUL_ALG_CONFIG_ID, &alg, sizeof(alg)); size_t workspace_size, compressed_size; cusparseLtMatmulGetWorkspace(&handle, &alg_sel, &workspace_size); cusparseLtMatmulPlanInit(&handle, &plan, &matmul, &alg_sel, workspace_size);
如果稀疏矩陣尚未被其他進程剪枝，則可以在此時進行。不要忘記檢查稀疏模式的有效性，以確保可以使用稀疏張量核心來加速它。
cusparseLtSpMMAPrune(&handle, &matmul, dA, dA, CUSPARSELT_PRUNE_SPMMA_TILE, stream); // checking the correctness int is_valid = 0; cusparseLtSpMMAPruneCheck(&handle, &matmul, dA, &is_valid, stream); if (is_valid != 0) { std::printf("!!! The matrix does not conform to the SpMMA sparsity pattern. " “cusparseLtMatmul does not provide correct results\n”); return EXIT_FAILURE; }
現在已將矩陣A剪枝為2：4稀疏度，可以將其壓縮到大約原始大小的一半。與實際的矩陣乘法（小于5％）相比，該步驟的執行時間可以忽略不計。
cusparseLtSpMMACompressedSize(&handle, &plan, &compressed_size); cudaMalloc((void**) &dA_compressed, compressed_size); cusparseLtSpMMACompress(&handle, &plan, dA, dA_compressed, stream);
設置完成后，執行matmul操作。cusparseLtMatmul使用不同的B矩陣可以多次重復調用 。只需設置一次稀疏矩陣。對于A矩陣值更改的用例，cusparseLtSpMMACompress必須再次調用該例程以設置稀疏矩陣的數據結構。
void* d_workspace = nullptr; int num_streams = 0; cudaStream_t* streams = nullptr; cusparseLtMatmul(&handle, &plan, &alpha, dA_compressed, dB, &beta, dC, dD, d_workspace, streams, num_streams) )
最后，通過破壞matmul計劃和cuSPARSELt句柄來清理已使用的內存。
cusparseLtMatmulPlanDestroy（＆plan）;
cusparseLtDestroy（＆handle）;

cuSPARSELt
通過cuSPARSELt庫，可以輕松利用NVIDIA Sparse Tensor Core運算，從而在不降低網絡準確性的情況下，顯著提高了用于深度學習應用程序的矩陣矩陣乘法的性能。該庫還提供了用于矩陣壓縮，剪枝和性能自動調整的實用程序。簡而言之，與普通的密集數學方法相比，cuSPARSELt減少了計算，功耗，執行時間和內存存儲。

總結

以上是生活随笔為你收集整理的cuSPARSELt开发NVIDIA Ampere结构化稀疏性的全部內容，希望文章能夠幫你解決所遇到的問題。

如果覺得生活随笔網站內容還不錯，歡迎將生活随笔推薦給好友。