PaddlePaddle
diff --git a/‎cmake/hip.cmake
Copy file name to clipboard
+5-2Lines changed: 5 additions & 2 deletions b/‎cmake/hip.cmake
Copy file name to clipboard
+5-2Lines changed: 5 additions & 2 deletions
diff --git a/‎paddle/common/flags.cc
Copy file name to clipboard
+1-1Lines changed: 1 addition & 1 deletion b/‎paddle/common/flags.cc
Copy file name to clipboard
+1-1Lines changed: 1 addition & 1 deletion
diff --git a/‎paddle/fluid/inference/api/resource_manager.cc
Copy file name to clipboard
+4Lines changed: 4 additions & 0 deletions b/‎paddle/fluid/inference/api/resource_manager.cc
Copy file name to clipboard
+4Lines changed: 4 additions & 0 deletions
diff --git a/‎paddle/fluid/operators/fused/CMakeLists.txt
Copy file name to clipboard
+3Lines changed: 3 additions & 0 deletions b/‎paddle/fluid/operators/fused/CMakeLists.txt
Copy file name to clipboard
+3Lines changed: 3 additions & 0 deletions
diff --git a/‎paddle/phi/backends/dynload/CMakeLists.txt
Copy file name to clipboard
+1Lines changed: 1 addition & 0 deletions b/‎paddle/phi/backends/dynload/CMakeLists.txt
Copy file name to clipboard
+1Lines changed: 1 addition & 0 deletions
diff --git a/‎paddle/phi/backends/dynload/dynamic_loader.cc
Copy file name to clipboard
+2Lines changed: 2 additions & 0 deletions b/‎paddle/phi/backends/dynload/dynamic_loader.cc
Copy file name to clipboard
+2Lines changed: 2 additions & 0 deletions
diff --git a/‎paddle/phi/backends/dynload/hipblasLt.cc
Copy file name to clipboard
+24Lines changed: 24 additions & 0 deletions b/‎paddle/phi/backends/dynload/hipblasLt.cc
Copy file name to clipboard
+24Lines changed: 24 additions & 0 deletions
diff --git a/‎paddle/phi/backends/dynload/hipblasLt.h
Copy file name to clipboard
+78Lines changed: 78 additions & 0 deletions b/‎paddle/phi/backends/dynload/hipblasLt.h
Copy file name to clipboard
+78Lines changed: 78 additions & 0 deletions
diff --git a/‎paddle/phi/backends/gpu/forwards.h
Copy file name to clipboard
+3Lines changed: 3 additions & 0 deletions b/‎paddle/phi/backends/gpu/forwards.h
Copy file name to clipboard
+3Lines changed: 3 additions & 0 deletions
diff --git a/‎paddle/phi/backends/gpu/gpu_decls.h
Copy file name to clipboard
+1-3Lines changed: 1 addition & 3 deletions b/‎paddle/phi/backends/gpu/gpu_decls.h
Copy file name to clipboard
+1-3Lines changed: 1 addition & 3 deletions
@@ -75,6 +75,7 @@ endmacro()
 
 find_package_and_include(miopen)
 find_package_and_include(rocblas)
+find_package_and_include(hipblaslt)
 find_package_and_include(hiprand)
 find_package_and_include(rocrand)
 find_package_and_include(rccl)
@@ -91,9 +92,10 @@ endif()
 
 # set CXX flags for HIP
 set(CMAKE_C_FLAGS
-    "${CMAKE_C_FLAGS} -D__HIP_PLATFORM_HCC__ -DROCM_NO_WRAPPER_HEADER_WARNING")
+    "${CMAKE_C_FLAGS} -D__HIP_PLATFORM_HCC__ -D__HIP_PLATFORM_AMD__ -DROCM_NO_WRAPPER_HEADER_WARNING"
+)
 set(CMAKE_CXX_FLAGS
-    "${CMAKE_CXX_FLAGS} -D__HIP_PLATFORM_HCC__ -DROCM_NO_WRAPPER_HEADER_WARNING"
+    "${CMAKE_CXX_FLAGS} -D__HIP_PLATFORM_HCC__ -D__HIP_PLATFORM_AMD__ -DROCM_NO_WRAPPER_HEADER_WARNING"
 )
 set(CMAKE_CXX_FLAGS
     "${CMAKE_CXX_FLAGS} -DTHRUST_DEVICE_SYSTEM=THRUST_DEVICE_SYSTEM_HIP")
@@ -102,6 +104,7 @@ set(THRUST_DEVICE_SYSTEM THRUST_DEVICE_SYSTEM_HIP)
 # define HIP_CXX_FLAGS
 list(APPEND HIP_CXX_FLAGS -fPIC)
 list(APPEND HIP_CXX_FLAGS -D__HIP_PLATFORM_HCC__=1)
+list(APPEND HIP_CXX_FLAGS -D__HIP_PLATFORM_AMD__=1)
 # Note(qili93): HIP has compile conflicts of float16.h as platform::float16 overload std::is_floating_point and std::is_integer
 list(APPEND HIP_CXX_FLAGS -D__HIP_NO_HALF_CONVERSIONS__=1)
 list(APPEND HIP_CXX_FLAGS -DROCM_NO_WRAPPER_HEADER_WARNING)
 
@@ -176,7 +176,7 @@ PHI_DEFINE_EXPORTED_string(
     "share-memory only.");
 #endif
 
-#if defined(PADDLE_WITH_CUDA)
+#if defined(PADDLE_WITH_CUDA) || defined(PADDLE_WITH_HIP)
 /**
  * CUDA related FLAG
  * Name: FLAGS_cublaslt_exhaustive_search_times
 
@@ -222,7 +222,11 @@ void GPUContextResource::DestroyBlasHandle() {
 }
 
 void GPUContextResource::InitBlasLtHandle() {
+#ifdef PADDLE_WITH_HIP
+  phi::InitBlasLtHandle(reinterpret_cast<void**>(&blaslt_handle_));
+#else  // PADDLE_WITH_CUDA
   phi::InitBlasLtHandle(&blaslt_handle_);
+#endif
 }
 
 void GPUContextResource::DestroyBlasLtHandle() {
 
@@ -48,4 +48,7 @@ if(WITH_GPU OR WITH_ROCM)
   if(CUDA_VERSION GREATER_EQUAL 11.6)
     op_library(fused_gemm_epilogue_op)
   endif()
+  if(WITH_ROCM)
+    op_library(fused_gemm_epilogue_op)
+  endif()
 endif()
@@ -25,6 +25,7 @@ if(WITH_ROCM)
     APPEND
     HIP_SRCS
     rocblas.cc
+    hipblasLt.cc
     miopen.cc
     hiprand.cc
     hipfft.cc
 
@@ -419,6 +419,8 @@ void* GetCublasLtDsoHandle() {
   }
 #elif !defined(__linux__) && defined(PADDLE_WITH_CUDA) && CUDA_VERSION >= 10010
   return GetDsoHandleFromSearchPath(FLAGS_cuda_dir, "libcublasLt.so");
+#elif defined(PADDLE_WITH_HIP)
+  return GetDsoHandleFromSearchPath(FLAGS_rocm_dir, "libhipblaslt.so");
 #else
   std::string warning_msg(
       "Your CUDA_VERSION less 10.1, not support CublasLt. "
 
@@ -0,0 +1,24 @@
+/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved.
+Licensed under the Apache License, Version 2.0 (the "License");
+you may not use this file except in compliance with the License.
+You may obtain a copy of the License at
+    https://www.apache.org/licenses/LICENSE-2.0
+Unless required by applicable law or agreed to in writing, software
+distributed under the License is distributed on an "AS IS" BASIS,
+WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+See the License for the specific language governing permissions and
+limitations under the License. */
+
+#include "paddle/phi/backends/dynload/hipblasLt.h"
+
+namespace phi {
+namespace dynload {
+std::once_flag hipblasLt_dso_flag;
+void *hipblasLt_dso_handle = nullptr;
+
+#define DEFINE_WRAP(__name) DynLoad__##__name __name
+
+HIPBLASLT_BLAS_ROUTINE_EACH(DEFINE_WRAP);
+
+}  // namespace dynload
+}  // namespace phi
@@ -0,0 +1,78 @@
+/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved.
+Copyright (c) 2022 NVIDIA Authors. All Rights Reserved.
+Licensed under the Apache License, Version 2.0 (the "License");
+you may not use this file except in compliance with the License.
+You may obtain a copy of the License at
+    https://www.apache.org/licenses/LICENSE-2.0
+Unless required by applicable law or agreed to in writing, software
+distributed under the License is distributed on an "AS IS" BASIS,
+WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+See the License for the specific language governing permissions and
+limitations under the License. */
+
+#pragma once
+
+#include <hip/hip_runtime.h>
+#include <hipblaslt/hipblaslt.h>
+
+#include <mutex>  // NOLINT
+#include <type_traits>
+
+#include "paddle/phi/backends/dynload/dynamic_loader.h"
+#include "paddle/phi/common/port.h"
+
+namespace phi {
+namespace dynload {
+
+extern std::once_flag hipblasLt_dso_flag;
+extern void *hipblasLt_dso_handle;
+
+/**
+ * The following macro definition can generate structs
+ * (for each function) to dynamic load hipblasLt routine
+ * via operator overloading.
+ *
+ * note: default dynamic linked libs
+ */
+#define DECLARE_DYNAMIC_LOAD_HIPBLASLT_WRAP(__name)                         \
+  struct DynLoad__##__name {                                                \
+    template <typename... Args>                                             \
+    inline auto operator()(Args... args) -> DECLARE_TYPE(__name, args...) { \
+      using hipblasLt_func =                                                \
+          decltype(::__name(std::declval<Args>()...)) (*)(Args...);         \
+      std::call_once(hipblasLt_dso_flag, []() {                             \
+        hipblasLt_dso_handle = phi::dynload::GetCublasLtDsoHandle();        \
+      });                                                                   \
+      static void *p_##__name = dlsym(hipblasLt_dso_handle, #__name);       \
+      return reinterpret_cast<hipblasLt_func>(p_##__name)(args...);         \
+    }                                                                       \
+  };                                                                        \
+  extern DynLoad__##__name __name
+
+#define HIPBLASLT_BLAS_ROUTINE_EACH(__macro)      \
+  __macro(hipblasLtCreate);                       \
+  __macro(hipblasLtDestroy);                      \
+  __macro(hipblasLtMatmul);                       \
+  __macro(hipblasLtMatmulDescCreate);             \
+  __macro(hipblasLtMatmulDescDestroy);            \
+  __macro(hipblasLtMatmulDescSetAttribute);       \
+  __macro(hipblasLtMatmulDescGetAttribute);       \
+  __macro(hipblasLtMatrixLayoutCreate);           \
+  __macro(hipblasLtMatrixLayoutDestroy);          \
+  __macro(hipblasLtMatrixLayoutSetAttribute);     \
+  __macro(hipblasLtMatrixLayoutGetAttribute);     \
+  __macro(hipblasLtMatmulPreferenceCreate);       \
+  __macro(hipblasLtMatmulPreferenceDestroy);      \
+  __macro(hipblasLtMatmulPreferenceSetAttribute); \
+  __macro(hipblasLtMatmulPreferenceGetAttribute); \
+  __macro(hipblasLtMatmulAlgoGetHeuristic);       \
+  __macro(hipblasLtMatrixTransform);              \
+  __macro(hipblasLtMatrixTransformDescCreate);    \
+  __macro(hipblasLtMatrixTransformDescDestroy);   \
+  __macro(hipblasLtMatrixTransformDescSetAttribute);
+
+HIPBLASLT_BLAS_ROUTINE_EACH(DECLARE_DYNAMIC_LOAD_HIPBLASLT_WRAP)
+
+#undef DECLARE_DYNAMIC_LOAD_HIPBLASLT_WRAP
+}  // namespace dynload
+}  // namespace phi
@@ -103,6 +103,9 @@ using miopenAllocatorFunction = void *(*)(void *context, size_t sizeBytes);
 // Forward declaration of rocBLAS types.
 using rocblas_handle = struct _rocblas_handle *;
 
+// Forward declaration of hipblaslt types.
+using hipblasLtHandle_t = void *;
+
 // Forward declaration of hipfft types.
 using hipfftHandle = struct hipfftHandle_t *;
 
 
@@ -60,9 +60,7 @@ DECLARE_TYPE_FOR_GPU(dnnHandle_t, cudnnHandle_t, miopenHandle_t);
 
 DECLARE_TYPE_FOR_GPU(blasHandle_t, cublasHandle_t, rocblas_handle);
 
-// TODO(Ming Huang): Since there is no blasLt handler,
-// use rocblas_handle for workround.
-DECLARE_TYPE_FOR_GPU(blasLtHandle_t, cublasLtHandle_t, rocblas_handle);
+DECLARE_TYPE_FOR_GPU(blasLtHandle_t, cublasLtHandle_t, hipblasLtHandle_t);
 
 DECLARE_TYPE_FOR_GPU(solverHandle_t, cusolverDnHandle_t, rocsolver_handle);
Original file line number	Diff line number	Diff line change
`@@ -222,7 +222,11 @@ void GPUContextResource::DestroyBlasHandle() {`
`222`	`222`	`}`
`223`	`223`
`224`	`224`	`void GPUContextResource::InitBlasLtHandle() {`
	`225`	`+#ifdef PADDLE_WITH_HIP`
	`226`	`+ phi::InitBlasLtHandle(reinterpret_cast<void**>(&blaslt_handle_));`
	`227`	`+#else // PADDLE_WITH_CUDA`
`225`	`228`	`phi::InitBlasLtHandle(&blaslt_handle_);`
	`229`	`+#endif`
`226`	`230`	`}`
`227`	`231`
`228`	`232`	`void GPUContextResource::DestroyBlasLtHandle() {`
Original file line number	Diff line number	Diff line change
`@@ -419,6 +419,8 @@ void* GetCublasLtDsoHandle() {`
`419`	`419`	`}`
`420`	`420`	`#elif !defined(__linux__) && defined(PADDLE_WITH_CUDA) && CUDA_VERSION >= 10010`
`421`	`421`	`return GetDsoHandleFromSearchPath(FLAGS_cuda_dir, "libcublasLt.so");`
	`422`	`+#elif defined(PADDLE_WITH_HIP)`
	`423`	`+ return GetDsoHandleFromSearchPath(FLAGS_rocm_dir, "libhipblaslt.so");`
`422`	`424`	`#else`
`423`	`425`	`std::string warning_msg(`
`424`	`426`	`"Your CUDA_VERSION less 10.1, not support CublasLt. "`