[DCU]Fix build error on rocm5.7 (PaddlePaddle#63738)

* [ROCM]Add patch for warprnnt/warpctc

* add rocm_path to cmake_prefix_path
* rm ambiguous apt because rocm has implemented

Signed-off-by: jiajuku <[email protected]>

* [ROCM] fix build error on rocm5.7

* change expression of Xcompiler since clang17 only support
"-Xcompiler xxx" instead of "-Xcompiler=xxx"
* clang17 check warning mre strictly

Signed-off-by: jiajuku <[email protected]>

* [ROCM] add dependent rocm lib

Signed-off-by: jiajuku <[email protected]>

* [ROCM]fix build warning

rocm5 requires the include path: $ROCM_PATH/include

Signed-off-by: jiajuku <[email protected]>

* clean code style issue

Signed-off-by: jiajuku <[email protected]>

* Revert "[ROCM] fix build error on rocm5.7"

This reverts commit 2b3bb69.

* [ROCm]:support rocm 57

* change build option "-Xcompiler" expression, since clang17 reuqire
  "-Xcompiler xxx" instead of "-Xcompiler=xxx"
* remove Werror from HIP_HCC_FLAGS since clang17 check more strictly.

Signed-off-by: onepick <[email protected]>

* Fix build error Xcompiler expression for clang17

Signed-off-by: onepick <[email protected]>

* Fix build warning

* sign-compare
* bitwise-instead-of-logical
* unknown-warning-option
* unused-lambda-capture
* unused-variable
* unused-but-set-variable
* reorder-ctor
* deprecated-copy-with-user-provided-copy
* unused-local-typedef
* missing-braces
* sometimes-uninitialized

Signed-off-by: onepick <[email protected]>

---------

Signed-off-by: jiajuku <[email protected]>
Signed-off-by: onepick <[email protected]>
Signed-off-by: onepick <[email protected]>
Co-authored-by: onepick <[email protected]>
Co-authored-by: onepick <[email protected]>

Author: 3 people

cmake/external/warpctc.cmake

@@ -49,6 +49,13 @@ if(NOT WIN32 AND WITH_GPU)
   endif()
 endif()
 
+if(WITH_ROCM)
+  set(WARPCTC_PATHCH_ROCM_COMMAND
+      patch -p1 <
+      ${PADDLE_SOURCE_DIR}/patches/warpctc/CMakeLists.txt.rocm.patch && patch
+      -p1 < ${PADDLE_SOURCE_DIR}/patches/warpctc/devicetypes.cuh.patch)
+endif()
+
 set(WARPCTC_INCLUDE_DIR
     "${WARPCTC_INSTALL_DIR}/include"
     CACHE PATH "Warp-ctc Directory" FORCE)
@@ -100,7 +107,10 @@ ExternalProject_Add(
   SOURCE_DIR ${SOURCE_DIR}
   PREFIX ${WARPCTC_PREFIX_DIR}
   UPDATE_COMMAND ""
-  PATCH_COMMAND ${WARPCTC_PATCH_COMMAND} ${WARPCTC_PATCH_CUDA_COMMAND}
+  PATCH_COMMAND
+  COMMAND ${WARPCTC_PATCH_COMMAND}
+  COMMAND ${WARPCTC_PATCH_CUDA_COMMAND}
+  COMMAND ${WARPCTC_PATHCH_ROCM_COMMAND}
   #BUILD_ALWAYS    1
   CMAKE_ARGS -DCMAKE_CXX_COMPILER=${CMAKE_CXX_COMPILER}
              -DCMAKE_C_COMPILER=${CMAKE_C_COMPILER}

cmake/external/warprnnt.cmake

@@ -35,7 +35,11 @@ else()
       ${SOURCE_DIR} <
       ${PADDLE_SOURCE_DIR}/patches/warprnnt/CMakeLists.txt.cuda.patch)
 endif()
-
+if(WITH_ROCM)
+  set(WARPRNNT_PATCH_ROCM_COMMAND
+      patch -p1 <
+      ${PADDLE_SOURCE_DIR}/patches/warprnnt/CMakeLists.txt.rocm.patch)
+endif()
 if(NOT WIN32 AND WITH_GPU)
   if(${CMAKE_CUDA_COMPILER_VERSION} LESS 12.0 AND ${CMAKE_CXX_COMPILER_VERSION}
                                                   VERSION_GREATER 12.0)
@@ -99,7 +103,9 @@ ExternalProject_Add(
   SOURCE_DIR ${SOURCE_DIR}
   PREFIX ${WARPRNNT_PREFIX_DIR}
   UPDATE_COMMAND ""
-  PATCH_COMMAND ${WARPCTC_PATCH_CUDA_COMMAND}
+  PATCH_COMMAND
+  COMMAND ${WARPCTC_PATCH_CUDA_COMMAND}
+  COMMAND ${WARPRNNT_PATCH_ROCM_COMMAND}
   #BUILD_ALWAYS    1
   CMAKE_ARGS -DCMAKE_CXX_COMPILER=${CMAKE_CXX_COMPILER}
              -DCMAKE_C_COMPILER=${CMAKE_C_COMPILER}

cmake/flags.cmake

@@ -92,8 +92,13 @@ macro(safe_set_nvflag flag_name)
   check_c_compiler_flag(${flag_name} C_COMPILER_SUPPORT_FLAG_${safe_name})
   set(safe_name C_COMPILER_SUPPORT_FLAG_${safe_name})
   if(${safe_name})
-    set(SAFE_GPU_COMMON_FLAGS
-        "${SAFE_GPU_COMMON_FLAGS} -Xcompiler=\"${flag_name}\"")
+    if(WITH_ROCM)
+      set(SAFE_GPU_COMMON_FLAGS
+          "${SAFE_GPU_COMMON_FLAGS} -Xcompiler \"${flag_name}\"")
+    else()
+      set(SAFE_GPU_COMMON_FLAGS
+          "${SAFE_GPU_COMMON_FLAGS} -Xcompiler=\"${flag_name}\"")
+    endif()
   endif()
 endmacro()
 
@@ -279,6 +284,7 @@ endif()
 
 # Disable -Werror, otherwise the compile will fail for rocblas_gemm_ex
 if(WITH_ROCM)
+  string(REPLACE "-Werror" "-Wno-error" HIP_HIPCC_FLAGS ${HIP_HIPCC_FLAGS})
   string(REPLACE "-Werror" "-Wno-error" CMAKE_CXX_FLAGS ${CMAKE_CXX_FLAGS})
   string(REPLACE "-Werror" "-Wno-error" CMAKE_C_FLAGS ${CMAKE_C_FLAGS})
 endif()

cmake/hip.cmake

@@ -24,6 +24,7 @@ else()
       CACHE PATH "Path to which clang has been installed")
 endif()
 set(CMAKE_MODULE_PATH "${HIP_PATH}/cmake" ${CMAKE_MODULE_PATH})
+set(CMAKE_PREFIX_PATH "${ROCM_PATH}" ${CMAKE_PREFIX_PATH})
 
 find_package(HIP REQUIRED)
 include_directories(${ROCM_PATH}/include)
@@ -123,6 +124,17 @@ list(APPEND HIP_CXX_FLAGS -Wno-switch)
 list(APPEND HIP_CXX_FLAGS -Wno-literal-conversion)
 list(APPEND HIP_CXX_FLAGS -Wno-constant-conversion)
 list(APPEND HIP_CXX_FLAGS -Wno-defaulted-function-deleted)
+list(APPEND HIP_CXX_FLAGS -Wno-sign-compare)
+list(APPEND HIP_CXX_FLAGS -Wno-bitwise-instead-of-logical)
+list(APPEND HIP_CXX_FLAGS -Wno-unknown-warning-option)
+list(APPEND HIP_CXX_FLAGS -Wno-unused-lambda-capture)
+list(APPEND HIP_CXX_FLAGS -Wno-unused-variable)
+list(APPEND HIP_CXX_FLAGS -Wno-unused-but-set-variable)
+list(APPEND HIP_CXX_FLAGS -Wno-reorder-ctor)
+list(APPEND HIP_CXX_FLAGS -Wno-deprecated-copy-with-user-provided-copy)
+list(APPEND HIP_CXX_FLAGS -Wno-unused-local-typedef)
+list(APPEND HIP_CXX_FLAGS -Wno-missing-braces)
+list(APPEND HIP_CXX_FLAGS -Wno-sometimes-uninitialized)
 
 if(WITH_CINN)
   list(APPEND HIP_CXX_FLAGS -std=c++14)

paddle/fluid/framework/var_type_traits.h

@@ -37,7 +37,7 @@
 #ifdef PADDLE_WITH_HIP
 #include <miopen/miopen.h>
 #ifdef PADDLE_WITH_RCCL
-#include <rccl.h>
+#include <rccl/rccl.h>
 #endif
 #endif
 

paddle/fluid/memory/allocation/CMakeLists.txt

@@ -33,10 +33,13 @@ if(WITH_GPU OR WITH_ROCM)
   list(APPEND ALLOCATOR_DEPS cuda_device_guard gpu_info dynload_cuda)
 endif()
 
-if(WITH_GPU)
-  list(APPEND ALLOCATOR_DEPS phi common)
+if(WITH_ROCM)
+  list(APPEND ALLOCATOR_DEPS ${ROCM_HIPRTC_LIB})
 endif()
 
+if(WITH_GPU OR WITH_ROCM)
+  list(APPEND ALLOCATOR_DEPS phi common)
+endif()
 if(CUDA_VERSION VERSION_GREATER_EQUAL 10.2)
   list(APPEND ALLOCATOR_SRCS cuda_virtual_mem_allocator.cc)
 endif()

paddle/fluid/platform/dynload/rccl.h

@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 #pragma once
 
-#include <rccl.h>
+#include <rccl/rccl.h>
 
 #include <mutex>  // NOLINT
 

paddle/fluid/platform/dynload/rocblas.h

@@ -15,7 +15,7 @@ limitations under the License. */
 #pragma once
 
 #include <hip/hip_runtime.h>
-#include <rocblas.h>
+#include <rocblas/rocblas.h>
 
 #include <mutex>  // NOLINT
 #include <type_traits>

paddle/fluid/platform/enforce.h

@@ -39,9 +39,9 @@ limitations under the License. */
 #endif  // PADDLE_WITH_CUDA
 
 #ifdef PADDLE_WITH_HIP
-#include <hiprand.h>
+#include <hiprand/hiprand.h>
 #include <miopen/miopen.h>
-#include <rocblas.h>
+#include <rocblas/rocblas.h>
 #include <thrust/system/hip/error.h>
 #include <thrust/system_error.h>  // NOLINT
 #endif

paddle/phi/backends/dynload/hipfft.h

@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 #pragma once
 #ifdef PADDLE_WITH_HIP
-#include <hipfft.h>
+#include <hipfft/hipfft.h>
 
 #include <mutex>  // NOLINT
 

paddle/phi/backends/dynload/hiprand.h

@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 #pragma once
 
-#include <hiprand.h>
+#include <hiprand/hiprand.h>
 
 #include <mutex>  // NOLINT
 

paddle/phi/backends/dynload/rccl.h

@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 #pragma once
 
-#include <rccl.h>
+#include <rccl/rccl.h>
 
 #include <mutex>  // NOLINT
 

paddle/phi/backends/dynload/rocblas.h

@@ -15,7 +15,7 @@ limitations under the License. */
 #pragma once
 
 #include <hip/hip_runtime.h>
-#include <rocblas.h>
+#include <rocblas/rocblas.h>
 
 #include <mutex>  // NOLINT
 #include <type_traits>

paddle/phi/core/enforce.h

@@ -24,9 +24,9 @@ limitations under the License. */
 #endif  // PADDLE_WITH_CUDA
 
 #ifdef PADDLE_WITH_HIP
-#include <hiprand.h>
+#include <hiprand/hiprand.h>
 #include <miopen/miopen.h>
-#include <rocblas.h>
+#include <rocblas/rocblas.h>
 #include <thrust/system/hip/error.h>
 #include <thrust/system_error.h>  // NOLINT
 #endif

patches/warpctc/CMakeLists.txt.rocm.patch

@@ -0,0 +1,10 @@
+--- a/CMakeLists.txt
++++ b/CMakeLists.txt
+@@ -24,6 +24,7 @@ option(BUILD_SHARED "build warp-ctc shared library."  ON)
+ option(WITH_ROCM     "Compile PaddlePaddle with ROCM platform"             OFF)
+
+ if(WITH_ROCM)
++    list(APPEND CMAKE_PREFIX_PATH ${ROCM_PATH})
+     add_definitions(-DWARPCTC_WITH_HIP)
+     include(hip)
+ endif(WITH_ROCM)

patches/warpctc/devicetypes.cuh.patch

@@ -0,0 +1,13 @@
+--- a/include/contrib/moderngpu/include/device/devicetypes.cuh
++++ b/include/contrib/moderngpu/include/device/devicetypes.cuh
+@@ -207,10 +207,6 @@ MGPU_HOST_DEVICE int2& operator+=(int2& a, int2 b) {
+ MGPU_HOST_DEVICE int2 operator*(int2 a, int2 b) {
+ 	return make_int2(a.x * b.x, a.y * b.y);
+ }
+-MGPU_HOST_DEVICE int2& operator*=(int2& a, int2 b) {
+-	a = a * b;
+-	return a;
+-}
+
+ template<typename T>
+ MGPU_HOST_DEVICE T max(T a, T b) {

patches/warprnnt/CMakeLists.txt.rocm.patch

@@ -0,0 +1,10 @@
+--- a/CMakeLists.txt
++++ b/CMakeLists.txt
+@@ -39,6 +39,7 @@ option(BUILD_SHARED "build warp-rnnt shared library."  ON)
+ option(WITH_ROCM   "Compile PaddlePaddle with ROCM platform" OFF)
+
+ if(WITH_ROCM)
++    list(APPEND CMAKE_PREFIX_PATH ${ROCM_PATH})
+     add_definitions(-DWARPRNNT_WITH_HIP)
+     include(hip)
+ endif(WITH_ROCM)