Use out-of-place trmm functions if rocBLAS version >= 4.0.0

uxlfoundation · Feb 20, 2024 · 0e397a1 · 0e397a1
1 parent 6774231
commit 0e397a1
Show file tree

Hide file tree

Showing 3 changed files with 25 additions and 0 deletions.
diff --git a/cmake/FindrocBLAS.cmake b/cmake/FindrocBLAS.cmake
@@ -34,6 +34,7 @@ list(APPEND CMAKE_PREFIX_PATH
 
 find_package(HIP QUIET)
 find_package(rocblas REQUIRED)
+set(ROCBLAS_VERSION ${rocblas_VERSION})
 
 # this is work around to avoid duplication half creation in both HIP and SYCL
 add_compile_definitions(HIP_NO_HALF)
@@ -47,6 +48,8 @@ find_package_handle_standard_args(rocBLAS
       HIP_LIBRARIES
       ROCBLAS_INCLUDE_DIR
       ROCBLAS_LIBRARIES
+    VERSION_VAR
+      ROCBLAS_VERSION
 )
 # OPENCL_INCLUDE_DIR
 if(NOT TARGET ONEMKL::rocBLAS::rocBLAS)

diff --git a/src/blas/backends/rocblas/CMakeLists.txt b/src/blas/backends/rocblas/CMakeLists.txt
@@ -41,6 +41,10 @@ target_include_directories(${LIB_OBJ}
           ${ONEMKL_GENERATED_INCLUDE_PATH}
 )
 
+if (${ROCBLAS_VERSION} VERSION_GREATER_EQUAL "4.0")
+  target_compile_definitions(${LIB_OBJ} PRIVATE ROCBLAS_NO_LEGACY_TRMM)
+endif()
+
 if(NOT ${ONEMKL_SYCL_IMPLEMENTATION} STREQUAL "hipsycl")
     target_compile_options(${LIB_OBJ} PRIVATE ${ONEMKL_BUILD_COPT})
     target_compile_options(ONEMKL::SYCL::SYCL INTERFACE

diff --git a/src/blas/backends/rocblas/rocblas_level3.cpp b/src/blas/backends/rocblas/rocblas_level3.cpp
@@ -381,10 +381,19 @@ inline void trmm(Func func, sycl::queue &queue, side left_right, uplo upper_lowe
             auto a_ = sc.get_mem<rocDataType *>(a_acc);
             auto b_ = sc.get_mem<rocDataType *>(b_acc);
             rocblas_status err;
+
+// rocblas version 4.0.0 removed the legacy BLAS trmm implementation
+#ifdef ROCBLAS_NO_LEGACY_TRMM
             ROCBLAS_ERROR_FUNC_SYNC(func, err, handle, get_rocblas_side_mode(left_right),
+                                    get_rocblas_fill_mode(upper_lower),
+                                    get_rocblas_operation(trans), get_rocblas_diag_type(unit_diag),
+                                    m, n, (rocDataType *)&alpha, a_, lda, b_, ldb, b_, ldb);
+#else
+	    ROCBLAS_ERROR_FUNC_SYNC(func, err, handle, get_rocblas_side_mode(left_right),
                                     get_rocblas_fill_mode(upper_lower),
                                     get_rocblas_operation(trans), get_rocblas_diag_type(unit_diag),
                                     m, n, (rocDataType *)&alpha, a_, lda, b_, ldb);
+#endif
         });
     });
 }
@@ -805,10 +814,19 @@ inline sycl::event trmm(Func func, sycl::queue &queue, side left_right, uplo upp
             auto a_ = reinterpret_cast<const rocDataType *>(a);
             auto b_ = reinterpret_cast<rocDataType *>(b);
             rocblas_status err;
+
+// rocblas version 4.0.0 removed the legacy BLAS trmm implementation
+#ifdef ROCBLAS_NO_LEGACY_TRMM
+            ROCBLAS_ERROR_FUNC_SYNC(func, err, handle, get_rocblas_side_mode(left_right),
+                                    get_rocblas_fill_mode(upper_lower),
+                                    get_rocblas_operation(trans), get_rocblas_diag_type(unit_diag),
+                                    m, n, (rocDataType *)&alpha, a_, lda, b_, ldb, b_, ldb);
+#else
             ROCBLAS_ERROR_FUNC_SYNC(func, err, handle, get_rocblas_side_mode(left_right),
                                     get_rocblas_fill_mode(upper_lower),
                                     get_rocblas_operation(trans), get_rocblas_diag_type(unit_diag),
                                     m, n, (rocDataType *)&alpha, a_, lda, b_, ldb);
+#endif
         });
     });