Merge pull request #2634 from JuliaGPU/tb/ctk_12.8

Update to CUDA 12.8

Author: maleadt

.buildkite/pipeline.yml

@@ -66,6 +66,7 @@ steps:
         matrix:
           setup:
             cuda:
+              - "12.8"
               - "12.6"
               - "12.5"
               - "12.4"

Project.toml

@@ -51,9 +51,9 @@ AbstractFFTs = "0.4, 0.5, 1.0"
 Adapt = "4"
 BFloat16s = "0.2, 0.3, 0.4, 0.5"
 CEnum = "0.2, 0.3, 0.4, 0.5"
-CUDA_Driver_jll = "0.10"
+CUDA_Driver_jll = "0.12"
 CUDA_Runtime_Discovery = "0.3.3"
-CUDA_Runtime_jll = "0.15"
+CUDA_Runtime_jll = "0.16"
 ChainRulesCore = "1"
 Crayons = "4"
 DataFrames = "1"

lib/cublas/libcublasLt.jl

@@ -718,7 +718,11 @@ const cublasLtMatmulPreference_t = Ptr{cublasLtMatmulPreferenceOpaque_t}
     CUBLASLT_MATMUL_TILE_768x56 = 628
     CUBLASLT_MATMUL_TILE_768x72 = 629
     CUBLASLT_MATMUL_TILE_768x80 = 630
-    CUBLASLT_MATMUL_TILE_END = 631
+    CUBLASLT_MATMUL_TILE_256x512 = 631
+    CUBLASLT_MATMUL_TILE_256x1024 = 632
+    CUBLASLT_MATMUL_TILE_512x512 = 633
+    CUBLASLT_MATMUL_TILE_512x1024 = 634
+    CUBLASLT_MATMUL_TILE_END = 635
 end
 
 @cenum cublasLtMatmulStages_t::UInt32 begin
@@ -757,7 +761,8 @@ end
     CUBLASLT_MATMUL_STAGES_32xAUTO = 34
     CUBLASLT_MATMUL_STAGES_64xAUTO = 35
     CUBLASLT_MATMUL_STAGES_128xAUTO = 36
-    CUBLASLT_MATMUL_STAGES_END = 37
+    CUBLASLT_MATMUL_STAGES_256xAUTO = 37
+    CUBLASLT_MATMUL_STAGES_END = 38
 end
 
 @cenum cublasLtClusterShape_t::UInt32 begin
@@ -824,6 +829,13 @@ end
     CUBLASLT_MATMUL_INNER_SHAPE_END = 5
 end
 
+@cenum cublasLtMatmulMatrixScale_t::UInt32 begin
+    CUBLASLT_MATMUL_MATRIX_SCALE_SCALAR_32F = 0
+    CUBLASLT_MATMUL_MATRIX_SCALE_VEC16_UE4M3 = 1
+    CUBLASLT_MATMUL_MATRIX_SCALE_VEC32_UE8M0 = 2
+    CUBLASLT_MATMUL_MATRIX_SCALE_END = 3
+end
+
 @cenum cublasLtPointerMode_t::UInt32 begin
     CUBLASLT_POINTER_MODE_HOST = 0
     CUBLASLT_POINTER_MODE_DEVICE = 1
@@ -973,6 +985,13 @@ end
     CUBLASLT_MATMUL_DESC_ATOMIC_SYNC_NUM_CHUNKS_D_COLS = 28
     CUBLASLT_MATMUL_DESC_ATOMIC_SYNC_IN_COUNTERS_POINTER = 29
     CUBLASLT_MATMUL_DESC_ATOMIC_SYNC_OUT_COUNTERS_POINTER = 30
+    CUBLASLT_MATMUL_DESC_A_SCALE_MODE = 31
+    CUBLASLT_MATMUL_DESC_B_SCALE_MODE = 32
+    CUBLASLT_MATMUL_DESC_C_SCALE_MODE = 33
+    CUBLASLT_MATMUL_DESC_D_SCALE_MODE = 34
+    CUBLASLT_MATMUL_DESC_EPILOGUE_AUX_SCALE_MODE = 35
+    CUBLASLT_MATMUL_DESC_D_OUT_SCALE_POINTER = 36
+    CUBLASLT_MATMUL_DESC_D_OUT_SCALE_MODE = 37
 end
 
 @checked function cublasLtMatmulDescInit_internal(matmulDesc, size, computeType, scaleType)
@@ -1318,7 +1337,7 @@ end
     @gcsafe_ccall libcublasLt.cublasLtLoggerSetMask(mask::Cint)::cublasStatus_t
 end
 
-# no prototype is found for this function at cublasLt.h:2448:29, please use with caution
+# no prototype is found for this function at cublasLt.h:2507:29, please use with caution
 @checked function cublasLtLoggerForceDisable()
     initialize_context()
     @gcsafe_ccall libcublasLt.cublasLtLoggerForceDisable()::cublasStatus_t

lib/cudadrv/libcuda.jl

@@ -37,6 +37,7 @@ end
     CUDNN_STATUS_LICENSE_ERROR = 1005
     CUDNN_STATUS_RUNTIME_IN_PROGRESS = 1006
     CUDNN_STATUS_RUNTIME_FP_OVERFLOW = 1007
+    CUDNN_STATUS_SUBLIBRARY_LOADING_FAILED = 1008
     CUDNN_STATUS_BAD_PARAM = 2000
     CUDNN_STATUS_BAD_PARAM_NULL_POINTER = 2002
     CUDNN_STATUS_BAD_PARAM_MISALIGNED_POINTER = 2003
@@ -380,12 +381,15 @@ end
     CUDNN_ATTR_ENGINECFG_ENGINE = 300
     CUDNN_ATTR_ENGINECFG_INTERMEDIATE_INFO = 301
     CUDNN_ATTR_ENGINECFG_KNOB_CHOICES = 302
+    CUDNN_ATTR_ENGINECFG_WORKSPACE_SIZE = 303
+    CUDNN_ATTR_ENGINECFG_SHARED_MEMORY_USED = 304
     CUDNN_ATTR_EXECUTION_PLAN_HANDLE = 400
     CUDNN_ATTR_EXECUTION_PLAN_ENGINE_CONFIG = 401
     CUDNN_ATTR_EXECUTION_PLAN_WORKSPACE_SIZE = 402
     CUDNN_ATTR_EXECUTION_PLAN_COMPUTED_INTERMEDIATE_UIDS = 403
     CUDNN_ATTR_EXECUTION_PLAN_RUN_ONLY_INTERMEDIATE_UIDS = 404
     CUDNN_ATTR_EXECUTION_PLAN_JSON_REPRESENTATION = 405
+    CUDNN_ATTR_EXECUTION_PLAN_KERNEL_CACHE = 406
     CUDNN_ATTR_INTERMEDIATE_INFO_UNIQUE_ID = 500
     CUDNN_ATTR_INTERMEDIATE_INFO_SIZE = 501
     CUDNN_ATTR_INTERMEDIATE_INFO_DEPENDENT_DATA_UIDS = 502
@@ -444,6 +448,7 @@ end
     CUDNN_ATTR_OPERATIONGRAPH_HANDLE = 800
     CUDNN_ATTR_OPERATIONGRAPH_OPS = 801
     CUDNN_ATTR_OPERATIONGRAPH_ENGINE_GLOBAL_COUNT = 802
+    CUDNN_ATTR_OPERATIONGRAPH_IS_DYNAMIC_SHAPE_ENABLED = 803
     CUDNN_ATTR_TENSOR_BYTE_ALIGNMENT = 900
     CUDNN_ATTR_TENSOR_DATA_TYPE = 901
     CUDNN_ATTR_TENSOR_DIMENSIONS = 902
@@ -530,6 +535,10 @@ end
     CUDNN_ATTR_OPERATION_SIGNAL_VALUE = 1902
     CUDNN_ATTR_OPERATION_SIGNAL_XDESC = 1903
     CUDNN_ATTR_OPERATION_SIGNAL_YDESC = 1904
+    CUDNN_ATTR_OPERATION_PAGED_CACHE_LOAD_CONTAINER_DESC = 1950
+    CUDNN_ATTR_OPERATION_PAGED_CACHE_LOAD_YDESC = 1951
+    CUDNN_ATTR_OPERATION_PAGED_CACHE_LOAD_SEQUENCE_DESC = 1952
+    CUDNN_ATTR_OPERATION_PAGED_CACHE_LOAD_PAGE_TABLE_DESC = 1953
     CUDNN_ATTR_OPERATION_NORM_FWD_MODE = 2000
     CUDNN_ATTR_OPERATION_NORM_FWD_PHASE = 2001
     CUDNN_ATTR_OPERATION_NORM_FWD_XDESC = 2002
@@ -568,6 +577,7 @@ end
     CUDNN_ATTR_OPERATION_RNG_SEED = 2311
     CUDNN_ATTR_OPERATION_RNG_DESC = 2312
     CUDNN_ATTR_OPERATION_RNG_OFFSET_DESC = 2313
+    CUDNN_ATTR_KERNEL_CACHE_IS_ENGINECFG_KERNEL_CACHED = 2400
 end
 
 @cenum cudnnBackendAttributeType_t::UInt32 begin
@@ -638,6 +648,8 @@ end
     CUDNN_BACKEND_OPERATION_RESHAPE_DESCRIPTOR = 31
     CUDNN_BACKEND_RNG_DESCRIPTOR = 32
     CUDNN_BACKEND_OPERATION_RNG_DESCRIPTOR = 33
+    CUDNN_BACKEND_KERNEL_CACHE_DESCRIPTOR = 34
+    CUDNN_BACKEND_OPERATION_PAGED_CACHE_LOAD_DESCRIPTOR = 35
 end
 
 @cenum cudnnBackendNumericalNote_t::UInt32 begin
@@ -2831,7 +2843,7 @@ end
                                                            ctcLossDesc::cudnnCTCLossDescriptor_t,
                                                            probsDesc::cudnnTensorDescriptor_t,
                                                            gradientsDesc::cudnnTensorDescriptor_t,
-                                                           sizeInBytes::Ptr{Csize_t})::cudnnStatus_t
+                                                           sizeInBytes::Ref{Csize_t})::cudnnStatus_t
 end
 
 mutable struct cudnnConvolutionStruct end