Fix handling of hibBLASLt

Author: jakurzak

src/BatchedGemm.h

@@ -94,7 +94,8 @@ class BatchedGemm {
                              int batch_count,
                              void const* alpha,
                              void const* beta,
-                             int device_id = 0);
+                             int device_id,
+                             bool lt);
 
     static void makeDevices(std::string type_a_name,
                             std::string type_b_name,
@@ -107,6 +108,7 @@ class BatchedGemm {
                             int batch_count,
                             void const* alpha,
                             void const* beta,
+                            bool lt,
                             bool host_a,
                             bool host_b,
                             bool host_c,

src/BatchedGemm.hip.cpp

@@ -38,6 +38,9 @@
 /// \param[in] device_id
 ///     the number of the GPU executing the matrix multiplication
 ///
+/// \param[in] lt
+///     If true, indicates the use of hipBLASLt.
+///
 BatchedGemm*
 BatchedGemm::make(std::string type_a_name,
                   std::string type_b_name,
@@ -50,7 +53,8 @@ BatchedGemm::make(std::string type_a_name,
                   int batch_count,
                   void const* alpha,
                   void const* beta,
-                  int device_id)
+                  int device_id,
+                  bool lt)
 {
     hipblasOperation_t op_a = stringToOp(op_a_name);
     hipblasOperation_t op_b = stringToOp(op_b_name);
@@ -83,7 +87,8 @@ BatchedGemm::make(std::string type_a_name,
                                  batch_count,
                                  alpha, beta,
                                  operations(stringToType(type_c_name), m, n, k),
-                                 device_id);
+                                 device_id,
+                                 lt);
 }
 
 //------------------------------------------------------------------------------
@@ -93,6 +98,9 @@ BatchedGemm::make(std::string type_a_name,
 /// \remark
 ///     Not listing parameters common with BatchedGemm::make().
 ///
+/// \param[in] lt
+///     If true, indicates the use of hipBLASLt.
+///
 /// \param[in] host_a, host_b, host_c
 ///     If true, indicates that the array is stored in host memory.
 ///
@@ -117,6 +125,7 @@ BatchedGemm::makeDevices(std::string type_a_name,
                          int batch_count,
                          void const* alpha,
                          void const* beta,
+                         bool lt,
                          bool host_a,
                          bool host_b,
                          bool host_c,
@@ -179,6 +188,7 @@ BatchedGemm::makeDevices(std::string type_a_name,
                                   batch_count,
                                   alpha, beta,
                                   operations(stringToType(type_c_name), m, n, k),
-                                  device_id);
+                                  device_id,
+                                  lt);
     }
 }

src/DeviceBatchedGemm.h

@@ -50,7 +50,8 @@ class DeviceBatchedGemm: public BatchedGemm {
                       void const* alpha,
                       void const* beta,
                       double operations,
-                      int device_id = 0);
+                      int device_id,
+                      bool lt);
     ~DeviceBatchedGemm();
 
     /// Populates the batch with random data.
@@ -88,6 +89,7 @@ class DeviceBatchedGemm: public BatchedGemm {
     void runBatchedGemmLt();
 
     int device_id_; ///< the number of the device executing the operation
+    bool lt_;       ///< true if using hipBLASLt
 
     hipStream_t hip_stream_;                      ///< stream
     hipblasHandle_t hipblas_handle_;              ///< hipBLAS handle

src/DeviceBatchedGemm.hip.cpp

@@ -37,6 +37,9 @@
 /// \param[in] device_id
 ///     the number of the device executing the operation
 ///
+/// \param[in] device_id
+///     true if using hipBLASLt
+///
 DeviceBatchedGemm::DeviceBatchedGemm(TypeConstant compute_type,
                                      hipblasOperation_t op_a,
                                      hipblasOperation_t op_b,
@@ -47,14 +50,16 @@ DeviceBatchedGemm::DeviceBatchedGemm(TypeConstant compute_type,
                                      void const* alpha,
                                      void const* beta,
                                      double operations,
-                                     int device_id)
+                                     int device_id,
+                                     bool lt)
     : BatchedGemm(compute_type,
                   op_a, op_b,
                   a, b, c,
                   batch_count,
                   alpha, beta,
                   operations),
-      device_id_(device_id)
+      device_id_(device_id),
+      lt_(lt)
 {
     // Set device, create stream.
     HIP_CALL(hipSetDevice(device_id_));
@@ -64,12 +69,14 @@ DeviceBatchedGemm::DeviceBatchedGemm(TypeConstant compute_type,
     HIPBLAS_CALL(hipblasCreate(&hipblas_handle_));
     HIPBLAS_CALL(hipblasSetStream(hipblas_handle_, hip_stream_));
 
-    // Create hipBLASLt handle and matmul descriptor.
-    HIPBLASLT_CALL(hipblasLtCreate(&hipblaslt_handle_));
-    HIPBLASLT_CALL(hipblasLtMatmulDescCreate(
-        &hipblaslt_matmul_desc_,
-        compute_type.compute_,
-        c->type().hip_));
+    if (lt) {
+        // Create hipBLASLt handle and matmul descriptor.
+        HIPBLASLT_CALL(hipblasLtCreate(&hipblaslt_handle_));
+        HIPBLASLT_CALL(hipblasLtMatmulDescCreate(
+            &hipblaslt_matmul_desc_,
+            compute_type.compute_,
+            c->type().hip_));
+    }
 
     // Create hipRAND generator, assign stream.
     HIPRAND_CALL(hiprandCreateGenerator(&hiprand_generator_,
@@ -92,9 +99,11 @@ DeviceBatchedGemm::~DeviceBatchedGemm()
     (void)hipSetDevice(device_id_);
 
     // Destroy all the handles.
+    if (lt_) {
+        (void)hipblasLtMatmulDescDestroy(hipblaslt_matmul_desc_);
+        (void)hipblasLtDestroy(hipblaslt_handle_);
+    }
     (void)hiprandDestroyGenerator(hiprand_generator_);
-    (void)hipblasLtMatmulDescDestroy(hipblaslt_matmul_desc_);
-    (void)hipblasLtDestroy(hipblaslt_handle_);
     (void)hipblasDestroy(hipblas_handle_);
     (void)hipStreamDestroy(hip_stream_);
 

src/gemm.cpp

@@ -169,6 +169,7 @@ void run(int argc, char** argv)
                              std::atoi(argv[12]),  // ldc
                              std::atoi(argv[13]),  // batch count
                              alpha, beta,
+                             lt,                   // hipBLASLt?
                              host_a, host_b, host_c,
                              coherent_a, coherent_b, coherent_c,
                              shared_a, shared_b,