[API] fix bmm api not support big tensor on window (#73373)

ZhangX-21 · web-flow · commit 57d91535621b · 2025-06-17T11:05:18.000+08:00
diff --git a/paddle/phi/backends/dynload/cublas.h b/paddle/phi/backends/dynload/cublas.h
@@ -17,7 +17,7 @@ limitations under the License. */
 #include <cublasXt.h>
 #include <cublas_v2.h>
 #include <cuda.h>
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
 #include <cublas_api.h>
 #endif
 
@@ -141,7 +141,7 @@ CUBLAS_BLAS_ROUTINE_EACH_R3(DECLARE_DYNAMIC_LOAD_CUBLAS_WRAP)
 CUBLAS_BLAS_ROUTINE_EACH_R4(DECLARE_DYNAMIC_LOAD_CUBLAS_WRAP)
 #endif
 
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
 #define CUBLAS_BLAS_ROUTINE_EACH_R5(__macro) \
   __macro(cublasGemmStridedBatchedEx_64);    \
   __macro(cublasGemmEx_64);                  \
diff --git a/paddle/phi/kernels/funcs/blas/blas_impl.cu.h b/paddle/phi/kernels/funcs/blas/blas_impl.cu.h
@@ -154,7 +154,7 @@ struct CUBlas<float> {
 // Because the gcc 4.8 doesn't expand template parameter pack that
 // appears in a lambda-expression, I can not use template parameter pack
 // here.
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
     VLOG(5) << "use_tensor_op_math: "
             << (dev_ctx->tensor_core_available() ? "True" : "False");
     dev_ctx->TensorCoreCublasCallIfAvailable([&](cublasHandle_t handle) {
@@ -523,7 +523,7 @@ struct CUBlas<phi::dtype::float16> {
                          cudaDataType_t Ctype,
                          int64_t ldc,
                          cudaDataType_t computeType) {
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
     cublasGemmAlgo_t algo = CUBLAS_GEMM_DFALT;
     bool use_tensor_op_math = dev_ctx->tensor_core_available();
     if (use_tensor_op_math) {
@@ -786,7 +786,7 @@ struct CUBlas<phi::dtype::complex<float>> {
                          cudaDataType_t Ctype,
                          int64_t ldc,
                          cudaDataType_t computeType) {
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
     cublasGemmAlgo_t algo = CUBLAS_GEMM_DFALT;
     bool use_tensor_op_math = dev_ctx->tensor_core_available();
     if (use_tensor_op_math) {
@@ -1165,7 +1165,7 @@ struct CUBlas<phi::dtype::complex<double>> {
                          cudaDataType_t Ctype,
                          int64_t ldc,
                          cudaDataType_t computeType) {
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
     cublasGemmAlgo_t algo = CUBLAS_GEMM_DFALT;
     bool use_tensor_op_math = dev_ctx->tensor_core_available();
     if (use_tensor_op_math) {
@@ -1284,7 +1284,7 @@ void Blas<phi::GPUContext>::GEMM(CBLAS_TRANSPOSE transA,
   if (FLAGS_enable_cublas_tensor_op_math && std::is_same<T, float>::value) {
     auto &cuda_ctx = const_cast<phi::GPUContext &>(context_);
     if (M > INT_MAX_VALUE || N > INT_MAX_VALUE || K > INT_MAX_VALUE) {
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
       CUBlas<T>::GEMM_EX_64(&cuda_ctx,
                             cuTransB,
                             cuTransA,
@@ -1394,7 +1394,7 @@ inline void Blas<phi::GPUContext>::GEMM(CBLAS_TRANSPOSE transA,
   // using tensor cores in volta GPUs.
   auto &cuda_ctx = const_cast<phi::GPUContext &>(context_);
   if (M > INT_MAX_VALUE || N > INT_MAX_VALUE || K > INT_MAX_VALUE) {
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
     CUBlas<phi::dtype::float16>::GEMM_EX_64(&cuda_ctx,
                                             cuTransB,
                                             cuTransA,
@@ -1491,7 +1491,7 @@ void Blas<phi::GPUContext>::GEMM(CBLAS_TRANSPOSE transA,
   if (FLAGS_enable_cublas_tensor_op_math && std::is_same<T, float>::value) {
     auto &cuda_ctx = const_cast<phi::GPUContext &>(context_);
     if (M > INT_MAX_VALUE || N > INT_MAX_VALUE || K > INT_MAX_VALUE) {
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
       CUBlas<T>::GEMM_EX_64(&cuda_ctx,
                             cuTransB,
                             cuTransA,
@@ -1602,7 +1602,7 @@ inline void Blas<phi::GPUContext>::GEMM(CBLAS_TRANSPOSE transA,
   // input/output in fp16, computation in fp32, which can also be accelerated
   // using tensor cores in volta GPUs.
   if (M > INT_MAX_VALUE || N > INT_MAX_VALUE || K > INT_MAX_VALUE) {
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
     CUBlas<phi::dtype::float16>::GEMM_EX_64(&cuda_ctx,
                                             cuTransB,
                                             cuTransA,
@@ -1707,7 +1707,7 @@ inline void Blas<phi::GPUContext>::GEMM(CBLAS_TRANSPOSE transA,
   }
   VLOG(5) << "use_tensor_op_math: " << (use_tensor_op_math ? "True" : "False");
   if (M > INT_MAX_VALUE || N > INT_MAX_VALUE || K > INT_MAX_VALUE) {
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
     cublasComputeType_t migratedComputeType = CUBLAS_COMPUTE_32F;
     context_.TensorCoreCublasCallIfAvailable([&](cublasHandle_t handle) {
       PADDLE_ENFORCE_GPU_SUCCESS(
@@ -1807,7 +1807,7 @@ inline void Blas<phi::GPUContext>::GEMM(CBLAS_TRANSPOSE transA,
   }
   VLOG(5) << "use_tensor_op_math: " << (use_tensor_op_math ? "True" : "False");
   if (M > INT_MAX_VALUE || N > INT_MAX_VALUE || K > INT_MAX_VALUE) {
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
     cublasComputeType_t migratedComputeType = CUBLAS_COMPUTE_32F;
     context_.TensorCoreCublasCallIfAvailable([&](cublasHandle_t handle) {
       PADDLE_ENFORCE_GPU_SUCCESS(
@@ -1906,7 +1906,7 @@ inline void Blas<phi::GPUContext>::GEMM(CBLAS_TRANSPOSE transA,
 #endif
 
   if (M > INT_MAX_VALUE || N > INT_MAX_VALUE || K > INT_MAX_VALUE) {
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
     CUBlas<phi::dtype::complex<float>>::GEMM_EX_64(&cuda_ctx,
                                                    cuTransB,
                                                    cuTransA,
@@ -2015,7 +2015,7 @@ inline void Blas<phi::GPUContext>::GEMM(CBLAS_TRANSPOSE transA,
   // input/output in fp16, computation in fp32, which can also be accelerated
   // using tensor cores in volta GPUs.
   if (M > INT_MAX_VALUE || N > INT_MAX_VALUE || K > INT_MAX_VALUE) {
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
     CUBlas<phi::dtype::complex<double>>::GEMM_EX_64(&cuda_ctx,
                                                     cuTransB,
                                                     cuTransA,
@@ -2391,7 +2391,7 @@ void Blas<phi::GPUContext>::BatchedGEMM(CBLAS_TRANSPOSE transA,
 #endif
     }
     if (M > INT_MAX_VALUE || N > INT_MAX_VALUE || K > INT_MAX_VALUE) {
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
       context_.TensorCoreCublasCallIfAvailable([&](cublasHandle_t handle) {
         PADDLE_ENFORCE_GPU_SUCCESS(
             phi::dynload::cublasGemmStridedBatchedEx_64(handle,
@@ -2541,7 +2541,7 @@ void Blas<phi::GPUContext>::BatchedGEMM(CBLAS_TRANSPOSE transA,
 
     if (M > INT_MAX_VALUE || N > INT_MAX_VALUE || K > INT_MAX_VALUE ||
         batchCount > INT_MAX_VALUE) {
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
       context_.TensorCoreCublasCallIfAvailable([&](cublasHandle_t handle) {
         PADDLE_ENFORCE_GPU_SUCCESS(
             phi::dynload::cublasGemmStridedBatchedEx_64(handle,
@@ -2669,7 +2669,7 @@ inline void Blas<phi::GPUContext>::BatchedGEMM(CBLAS_TRANSPOSE transA,
   VLOG(5) << "use_tensor_op_math: " << (use_tensor_op_math ? "True" : "False");
   if (M > INT_MAX_VALUE || N > INT_MAX_VALUE || K > INT_MAX_VALUE ||
       batchCount > INT_MAX_VALUE) {
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
     context_.TensorCoreCublasCallIfAvailable([&](cublasHandle_t handle) {
       PADDLE_ENFORCE_GPU_SUCCESS(
           phi::dynload::cublasGemmStridedBatchedEx_64(handle,
@@ -2774,7 +2774,7 @@ inline void Blas<phi::GPUContext>::BatchedGEMM(CBLAS_TRANSPOSE transA,
   VLOG(5) << "use_tensor_op_math: " << (use_tensor_op_math ? "True" : "False");
   if (M > INT_MAX_VALUE || N > INT_MAX_VALUE || K > INT_MAX_VALUE ||
       batchCount > INT_MAX_VALUE) {
-#if CUDA_VERSION >= 12030
+#if CUDA_VERSION >= 12030 && defined(__linux__)
     context_.TensorCoreCublasCallIfAvailable([&](cublasHandle_t handle) {
       PADDLE_ENFORCE_GPU_SUCCESS(
           phi::dynload::cublasGemmStridedBatchedEx_64(handle,
