Merge branch 'amd-develop' into amd-master

ROCm · Oct 18, 2022 · bceeae0 · bceeae0
2 parents 610e377 + 2a09ae4
commit bceeae0
Show file tree

Hide file tree

Showing 189 changed files with 92,986 additions and 54,079 deletions.
diff --git a/CMakeLists.txt b/CMakeLists.txt
@@ -101,7 +101,7 @@ endif()
 
 ############################################################
 # require C++14
-add_compile_options(-std=c++14)
+add_compile_options(-std=c++17)
 
 ############################################################
 # OPTION - MIOpen Backend
@@ -221,10 +221,11 @@ if( DEFINED MIOPEN_OVERRIDE_HIP_VERSION_PATCH )
 endif()
 
 # Depend on Composable Kernels
-option(MIOPEN_USE_COMPOSABLEKERNEL "Enable MIOpen to use composable kernels for various operations" Off)
-if(MIOPEN_USE_COMPOSABLEKERNEL)
-find_package(composable_kernel 1.0.0 COMPONENTS device_operations host_tensor)
+option(MIOPEN_USE_COMPOSABLEKERNEL "Enable MIOpen to use composable kernels for various operations" On)
+if(MIOPEN_BACKEND_OPENCL)
+    set(MIOPEN_USE_COMPOSABLEKERNEL OFF)
 endif()
+message(STATUS "Enable Composable Kernels: ${MIOPEN_USE_COMPOSABLEKERNEL}")
 
 set_var_to_condition(MIOPEN_USE_COMGR_DEFAULT (NOT DEFINED MIOPEN_BACKEND_OPENCL) AND (NOT (MIOPEN_BACKEND STREQUAL "HIPNOGPU")))
 option(MIOPEN_USE_COMGR "Use comgr to build kernels instead of offline tools" ${MIOPEN_USE_COMGR_DEFAULT})
@@ -246,6 +247,9 @@ add_definitions("-DHIP_COMPILER_FLAGS=${HIP_COMPILER_FLAGS}")
 
 # HIP
 if( MIOPEN_BACKEND STREQUAL "HIP" OR MIOPEN_BACKEND STREQUAL "HIPOC" OR MIOPEN_BACKEND STREQUAL "HIPNOGPU")
+    if(MIOPEN_USE_COMPOSABLEKERNEL)
+        find_package(composable_kernel 1.0.0 COMPONENTS device_operations)
+    endif()
     if( MIOPEN_BACKEND STREQUAL "HIPNOGPU")
         set(MIOPEN_MODE_NOGPU 1)
     endif()
@@ -281,6 +285,8 @@ if( MIOPEN_BACKEND STREQUAL "HIP" OR MIOPEN_BACKEND STREQUAL "HIPOC" OR MIOPEN_B
         message(STATUS "Build without rocblas")
     endif()
 else()
+    #CK is only enabled when HIP backend is selected   
+    set(MIOPEN_USE_COMPOSABLEKERNEL Off)
     if(MIOPEN_USE_HIPRTC)
         message(FATAL_ERROR "HIPRTC cannot be used without HIP backend")
     endif()
@@ -323,17 +329,25 @@ set_var_to_condition(MIOPEN_USE_MLIR_DEFAULT NOT (NOT ${BUILD_SHARED_LIBS} AND $
 option(MIOPEN_USE_MLIR "Use MLIR compilation backend" ${MIOPEN_USE_MLIR_DEFAULT})
 
 if(MIOPEN_USE_MLIR)
-    # REQUIRED is not supported before cmake 3.18
-    find_library(LIBMLIRMIOPEN MLIRMIOpen REQUIRED)
     if(NOT ${BUILD_SHARED_LIBS} AND ${MIOPEN_USE_COMGR})
         message(FATAL_ERROR "Potential symbol conflict between mlir and comgr in static build")
     endif()
-
-    if(NOT LIBMLIRMIOPEN)
-        message(FATAL_ERROR "libMLIRMIOpen not found, please reinstall dependencies. \
-        Refer to https://github.com/ROCmSoftwarePlatform/MIOpen#installing-the-dependencies")
+    # Try to find package rocMLIR
+    # REQUIRED is omitted since we do not want cmake to abort if the package is not found
+    find_package(rocMLIR 1.0.0 CONFIG)
+    if(NOT rocMLIR_FOUND)
+        message(STATUS "Falling back to find library libMLIRMIOpen")
+        # Backward compatibility with ROCm 5.3
+        # If the rocMLIR package is not found, try to find the library libMLIRMIOpen directly
+        find_library(LIBMLIRMIOPEN MLIRMIOpen REQUIRED)
+        if(NOT LIBMLIRMIOPEN)
+            message(FATAL_ERROR "library libMLIRMIOpen not found, please reinstall dependencies. \
+            Refer to https://github.com/ROCmSoftwarePlatform/MIOpen#installing-the-dependencies")
+        else()
+            message(STATUS "Build with library libMLIRMIOpen: " ${LIBMLIRMIOPEN})
+        endif()
     else()
-        message(STATUS "Build with libMLIRMIOpen: " ${LIBMLIRMIOPEN})
+        message(STATUS "Build with rocMLIR::rockCompiler ${rocMLIR_VERSION}")
     endif()
 endif()
 
@@ -594,6 +608,9 @@ enable_clang_tidy(
         -modernize-use-equals-default
         -modernize-use-trailing-return-type
         -modernize-use-transparent-functors
+        -modernize-use-nodiscard
+        -modernize-concat-nested-namespaces
+        -modernize-unary-static-assert
         -performance-unnecessary-value-param
         -readability-braces-around-statements
         -readability-convert-member-functions-to-static
@@ -618,7 +635,23 @@ enable_clang_tidy(
         -readability-container-data-pointer
         -readability-identifier-length
         -readability-suspicious-call-argument
-
+        #TODO Code Quality WORKAROUND ROCm 5.3 && Ubuntu 22.04 && C++17 && cppcheck 2.9 update
+        -bugprone-use-after-move
+        -hicpp-invalid-access-moved
+        -modernize-use-nodiscard
+        -misc-confusable-identifiers
+        -modernize-unary-static-assert
+        -modernize-macro-to-enum
+        -modernize-concat-nested-namespaces
+        -readability-redundant-declaration
+        -readability-simplify-boolean-expr
+        -bugprone-branch-clone
+        -clang-diagnostic-deprecated
+        -hicpp-deprecated-headers
+        -hicpp-member-init
+        -performance-no-automatic-move
+        -clang-analyzer-cplusplus.NewDeleteLeaks
+        -modernize-deprecated-headers
         ${MIOPEN_TIDY_CHECKS}
     ${MIOPEN_TIDY_ERRORS}
     HEADER_FILTER
@@ -663,6 +696,30 @@ enable_cppcheck(
         unreadVariable:*src/composable_kernel/host/*/*
         unreadVariable:*src/composable_kernel/external/*/*
         unmatchedSuppression
+        #TODO Code Quality WORKAROUND ROCm 5.3 && Ubuntu 22.04 && C++17 && cppcheck 2.9 update
+        ctuOneDefinitionRuleViolation:*test/*
+        ctuOneDefinitionRuleViolation:*src/composable_kernel/composable_kernel/*/*
+        ctuPointerArith:*test/*
+        constParameter
+        constVariable
+        variableScope
+        missingReturn
+        cstyleCast
+        unknownMacro
+        uselessCallsSubstr
+        uninitMemberVar
+        overlappingWriteUnion
+        operatorEqVarError
+        returnTempReference
+        objectIndex
+        integerOverflowCond
+        rethrowNoCurrentException
+        mismatchingContainers
+        unreadVariable
+        CastIntegerToAddressAtReturn
+        knownConditionTrueFalse
+        shadowFunction
+        moduloofone
     FORCE
     SOURCES
         addkernels/

diff --git a/Dockerfile b/Dockerfile
@@ -1,4 +1,5 @@
 FROM ubuntu:20.04 as miopen
+ARG DEBIAN_FRONTEND=noninteractive
 
 ARG USE_MLIR="OFF"
 
@@ -12,57 +13,49 @@ DEBIAN_FRONTEND=noninteractive apt-get install -y --allow-unauthenticated \
     ca-certificates \
     curl \
     libnuma-dev \
-    gnupg \
+    gnupg2 \
     wget
 
 #Add gpg keys
+ENV APT_KEY_DONT_WARN_ON_DANGEROUS_USAGE=DontWarn
 RUN apt-key adv --keyserver hkp://keyserver.ubuntu.com:80 --recv-keys 9386B48A1A693C5C && \
     wget -q -O - https://repo.radeon.com/rocm/rocm.gpg.key | apt-key add -
 
+RUN wget https://repo.radeon.com/amdgpu-install/5.3/ubuntu/focal/amdgpu-install_5.3.50300-1_all.deb  --no-check-certificate
+RUN apt-get update && \
+DEBIAN_FRONTEND=noninteractive apt-get install -y --allow-unauthenticated \
+    ./amdgpu-install_5.3.50300-1_all.deb
+
 # Add rocm repository
 # Note: The ROCm version with $USE_MLIR should keep in sync with default ROCm version
 # unless MLIR library is incompatible with current ROCm.
-RUN export ROCM_APT_VER=.apt_5.2.3;\
+RUN export ROCM_APT_VER=5.3;\
 echo $ROCM_APT_VER &&\
 sh -c 'echo deb [arch=amd64 trusted=yes] http://repo.radeon.com/rocm/apt/$ROCM_APT_VER/ ubuntu main > /etc/apt/sources.list.d/rocm.list'
 RUN sh -c "echo deb http://mirrors.kernel.org/ubuntu focal main universe | tee -a /etc/apt/sources.list"
 
+RUN amdgpu-install -y --usecase=rocm --no-dkms
+
 # Install dependencies
 RUN apt-get update && \
 DEBIAN_FRONTEND=noninteractive apt-get install -y --allow-unauthenticated \
     build-essential \
     cmake \
-    comgr \
-    clang-format-10 \
+    clang-format-12 \
     doxygen \
-    g++ \
     gdb \
     git \
-    hip-rocclr \
     lcov \
-    libelf-dev \
     libncurses5-dev \
-    libpthread-stubs0-dev \
     llvm-amdgpu \
     miopengemm \
     pkg-config \
-    python \
-    python3 \
-    python-dev \
     python3-dev \
     python3-pip \
-    python3-distutils \
     python3-venv \
-    software-properties-common \
-    rocm-dev \
-    rocm-device-libs \
-    rocm-opencl \
-    rocm-opencl-dev \
     rocblas \
     rpm \
-    zlib1g-dev \
-    kmod && \
-    apt-get remove -y rocm-cmake && \
+    software-properties-common && \
     apt-get clean && \
     rm -rf /var/lib/apt/lists/*
 
@@ -127,10 +120,4 @@ RUN if [ "$USE_TARGETID" = "ON" ] ; then export HIPCC_LINK_FLAGS_APPEND='-O3 -pa
 ARG MIOTENSILE_VER="default"
 RUN if [ "$USE_TARGETID" = "OFF" ] ; then echo "MIOpenTensile is not installed."; elif [ "$MIOTENSILE_VER" = "latest" ] ; then cget -p $PREFIX install ROCmSoftwarePlatform/MIOpenTensile@94a9047741d16a8eccd290131b78fb1aa69cdcdf; else cget -p $PREFIX install ROCmSoftwarePlatform/MIOpenTensile@94a9047741d16a8eccd290131b78fb1aa69cdcdf; fi
 
-ARG CK_COMMIT=91d8b7d67ae9dbf8a6e691ea3e17c0b9705c6ba7
-RUN rm -rf /tmp/ck* && mkdir /tmp/ck && wget -O ck.tar.gz https://www.github.com/rocmsoftwareplatform/composable_kernel/archive/${CK_COMMIT}.tar.gz -O /tmp/ck.tar.gz && \
-    tar zxvf /tmp/ck.tar.gz -C /tmp/ && mkdir /tmp/composable_kernel-${CK_COMMIT}/build &&\
-    cd /tmp/composable_kernel-${CK_COMMIT}/build && \
-    CXX=/opt/rocm/bin/hipcc cmake -DCMAKE_CXX_COMPILER_LAUNCHER="${COMPILER_LAUNCHER}" -DCMAKE_PREFIX_PATH=/opt/rocm -D CMAKE_CXX_FLAGS=" --offload-arch=gfx900 --offload-arch=gfx906 --offload-arch=gfx908 --offload-arch=gfx90a --offload-arch=gfx1030 -O3 " .. && \
-    make -j $(nproc) install && rm -rf /tmp/*
 RUN groupadd -f render
diff --git a/Jenkinsfile b/Jenkinsfile
@@ -278,7 +278,7 @@ def buildDocker(install_prefix)
     checkout scm
     def image_name = getDockerImageName(install_prefix)
     echo "Building Docker for ${image_name}"
-    def dockerArgs = "--build-arg BUILDKIT_INLINE_CACHE=1 --build-arg PREFIX=${install_prefix} --build-arg GPU_ARCH='gfx900;gfx906;gfx908;gfx90a;gfx90a:xnack-;gfx1030' --build-arg MIOTENSILE_VER='default' --build-arg USE_TARGETID='OFF' --build-arg USE_MLIR='ON' --build-arg USE_FIN='ON' "
+    def dockerArgs = "--build-arg BUILDKIT_INLINE_CACHE=1 --build-arg PREFIX=${install_prefix} --build-arg GPU_ARCH='gfx900;gfx906;gfx908;gfx90a:xnack-;gfx1030' --build-arg MIOTENSILE_VER='default' --build-arg USE_TARGETID='OFF' --build-arg USE_MLIR='ON' --build-arg USE_FIN='ON' "
     if(env.CCACHE_HOST)
     {
         def check_host = sh(script:"""(printf "PING\\r\\n";) | nc  -N ${env.CCACHE_HOST} 6379 """, returnStdout: true).trim()
@@ -471,7 +471,7 @@ pipeline {
                                 -o -iname \'*.cpp.in\' \
                                 -o -iname \'*.cl\' \
                                 | grep -v -E '(build/)|(install/)' \
-                                | xargs -n 1 -P 1 -I{} -t sh -c \'clang-format-10 -style=file {} | diff - {}\'"
+                                | xargs -n 1 -P 1 -I{} -t sh -c \'clang-format-12 -style=file {} | diff - {}\'"
                     }
                     steps{
                         buildHipClangJobAndReboot(setup_cmd: "", build_cmd: "", execute_cmd: execute_cmd, needs_gpu:false)
@@ -761,20 +761,6 @@ pipeline {
                         buildHipClangJobAndReboot(compiler: 'g++', setup_flags: Int8_flags, config_targets: Smoke_targets)
                     }
                 }
-                stage('Int8 Hip Debug CK gfx908') {
-                    when {
-                        beforeAgent true
-                        expression { params.TARGET_GFX908}
-                    }
-                    agent{ label rocmnode("gfx908") }
-                    // This stage should be removed when CK is enabled by default in MIOpen
-                    environment{
-                        Enable_CK = "-DMIOPEN_USE_COMPOSABLEKERNEL=On"
-                    }
-                    steps{
-                        buildHipClangJobAndReboot( build_type: 'debug', setup_flags: Enable_CK + Int8_flags , build_env: extra_log_env, test_flags: ' --verbose ')
-                    }
-                }
                 stage('Fp16 Hip Vega20') {
                     when {
                         beforeAgent true

diff --git a/cmake/EnableCompilerWarnings.cmake b/cmake/EnableCompilerWarnings.cmake
@@ -67,6 +67,7 @@ else()
             -Wunused
             -Wno-ignored-qualifiers
             -Wno-sign-compare
+            -Wno-deprecated
         )
         if (CMAKE_${COMPILER}_COMPILER_ID MATCHES "Clang")
             list(APPEND CMAKE_COMPILER_WARNINGS
@@ -92,6 +93,7 @@ else()
                 -Wno-unused-command-line-argument
                 -Wno-weak-vtables
                 -Wno-covered-switch-default
+                -Wno-unused-result
             )
         else()
             if (CMAKE_${COMPILER}_COMPILER_ID MATCHES "GNU" AND ${COMPILER} MATCHES "CXX")

diff --git a/cmake/googletest.cmake b/cmake/googletest.cmake
@@ -20,6 +20,7 @@ list(APPEND GTEST_CMAKE_CXX_FLAGS
      -Wno-unused-member-function
      -Wno-comma
      -Wno-old-style-cast
+     -Wno-deprecated
 )
 message(STATUS "Suppressing googltest warnings with flags: ${GTEST_CMAKE_CXX_FLAGS}")
 

diff --git a/dev-requirements.txt b/dev-requirements.txt
@@ -1,5 +1,3 @@
 ROCmSoftwarePlatform/rocm-recipes
-RadeonOpenCompute/rocm-cmake@04f694df2a8dc9d7e35fa4dee4ba5fa407ec04f8 --build
 -f requirements.txt
-# 1.90+
-danmar/cppcheck@dd05839a7e63ef04afd34711cb3e1e0ef742882f
+danmar/[email protected]
diff --git a/doc/requirements.txt b/doc/requirements.txt
@@ -24,8 +24,8 @@
 # 
 ################################################################################
 # Python 3.6+ required
-sphinx==4.1.2
-breathe==4.30.0
+sphinx==5.2.0
+breathe==4.34.0
 docutils<0.17 # sphinx-rtd-theme 0.5.2 requires docutils<0.17
 sphinx_rtd_theme==0.5.2
-myst-parser==0.15.1
+myst-parser==0.18.0
diff --git a/driver/activ_driver.hpp b/driver/activ_driver.hpp
@@ -250,7 +250,7 @@ int ActivationDriver<Tgpu, Tref>::AllocateBuffersAndCopy()
             double v = -alpha / beta;
             in[i]    = i % 2 ? RAN_GEN<Tgpu>(static_cast<Tgpu>((v + 0.005) / beta),
                                           static_cast<Tgpu>((v + 2.0) / beta))
-                          : RAN_GEN<Tgpu>(static_cast<Tgpu>((v - 2.0) / beta),
+                             : RAN_GEN<Tgpu>(static_cast<Tgpu>((v - 2.0) / beta),
                                           static_cast<Tgpu>((v - 0.005) / beta));
             break;
         }

diff --git a/driver/conv_driver.hpp b/driver/conv_driver.hpp
@@ -2175,7 +2175,7 @@ int ConvDriver<Tgpu, Tref>::RunForwardGPUReference()
     auto ref_solution_id = miopen::deref(convDesc).mode == miopenTranspose
                                ? miopen::solver::Id("ConvDirectNaiveConvBwd").Value()
                                : miopen::solver::Id("ConvDirectNaiveConvFwd").Value();
-    auto rc = miopenConvolutionForwardImmediate(handle,
+    auto rc              = miopenConvolutionForwardImmediate(handle,
                                                 weightTensor,
                                                 wei_dev->GetMem(),
                                                 inputTensor,
@@ -3140,7 +3140,7 @@ int ConvDriver<Tgpu, Tref>::RunBackwardDataGPUReference()
     auto ref_solution_id = miopen::deref(convDesc).mode == miopenTranspose
                                ? miopen::solver::Id("ConvDirectNaiveConvFwd").Value()
                                : miopen::solver::Id("ConvDirectNaiveConvBwd").Value();
-    auto rc = miopenConvolutionBackwardDataImmediate(handle,
+    auto rc              = miopenConvolutionBackwardDataImmediate(handle,
                                                      outputTensor,
                                                      dout_dev->GetMem(),
                                                      weightTensor,
@@ -3317,8 +3317,8 @@ int ConvDriver<Tgpu, Tref>::VerifyForward()
 
     const auto isInt8 = (data_type == miopenInt8 || data_type == miopenInt8x4);
     auto error        = is_fwd_run_failed ? std::numeric_limits<double>::max()
-                                   : (isInt8 ? miopen::rms_range(outhost.data, out_int8)
-                                             : miopen::rms_range(outhost.data, out.data));
+                                          : (isInt8 ? miopen::rms_range(outhost.data, out_int8)
+                                                    : miopen::rms_range(outhost.data, out.data));
 
     auto tolerance = GetDefaultTolerance();
     // iGemm's deviation is higher than other algorithms.

diff --git a/driver/conv_verify.hpp b/driver/conv_verify.hpp
@@ -119,7 +119,7 @@ void RunBackwardWeightsCPUVerify(std::vector<Tcheck_>& dwei_host,
         assert(stride_h == 1);
         assert(stride_w == 1);
 
-        std::fill(dwei_host.begin(), dwei_host.end(), (static_cast<Tcheck_>(0));
+        std::fill(dwei_host.begin(), dwei_host.end(), static_cast<Tcheck_>(0));
 
         int batch_sz              = out_n;
         int outputs               = out_c;
@@ -134,7 +134,7 @@ void RunBackwardWeightsCPUVerify(std::vector<Tcheck_>& dwei_host,
 
         int filter_size_w = wei_w;
         int filter_size_h = wei_h;
-        int kernel_sz    = filter_size_w * filter_size_h;
+        int kernel_sz     = filter_size_w * filter_size_h;
 
         int top_height = out_h;
         int top_width  = out_w;

diff --git a/driver/main.cpp b/driver/main.cpp
@@ -199,9 +199,9 @@ int main(int argc, char* argv[])
         return rc;
     }
 
-    int fargval = ((base_arg != "CBAInfer") && (base_arg != "CBAInferfp16"))
-                      ? drv->GetInputFlags().GetValueInt("forw")
-                      : 1;
+    int fargval       = ((base_arg != "CBAInfer") && (base_arg != "CBAInferfp16"))
+                            ? drv->GetInputFlags().GetValueInt("forw")
+                            : 1;
     bool bnFwdInVer   = (fargval == 2 && (base_arg == "bnorm"));
     bool verifyarg    = (drv->GetInputFlags().GetValueInt("verify") == 1);
     int cumulative_rc = 0; // Do not stop running tests in case of errors.