Squashed 'src/composable_kernel/' changes from 5781adf5c..31b403526

31b403526 Merge pull request #16 from ROCmSoftwarePlatform/develop b62bf8c3f Merge pull request #14 from ROCmSoftwarePlatform/miopen_downstream_init_integration ccc4a1d36 Merge pull request #8 from ROCmSoftwarePlatform/miopen_downstream_init_integration 67ad47e7c refactor 16effa767 refactor a91b68dfc DynamicBuffer, StaticBuffer, amd_buffer_load support customized value for invalid element 2cbabbba5 use int instead of index_t in kernel wrapper 0834bc763 compiler parameter use stream f2ac7832c make innner product compatiable on gfx900 4e57b30a6 rename c03045ce2 rename b2589957f update CK build script 2c48039d0 fix kernel filename d626dccc9 fix enum issue 643ebd4f3 tidy ddd49ec9e fix clang warning suppression 4f566c622 vector/scalar pointer cast use c-style pointer cast instead of reinterpret_cast 172036d72 add c-style pointer cast 76f313193 tidy d18428901 tidy f885c131d tidy 80120f0a0 tidy c3efeb5e2 tidy 56fc0842b tidy 54fba515b tidy e62bae7a4 tidy 24c872894 add tidy 61487e0a0 fix ae98b52ad remove online compilation from CK cb9542131 refactor 73ca97015 Merge commit '437cc595c6e206dfebb118985b5171bbc1e29eab' into composable_kernel_init_integration_v3 3b8664611 Merge pull request #7 from ROCmSoftwarePlatform/master d09ea4f4e Update develop (#5) 3d32ae940 add solver ConvIgemmFwdV6r1DlopsNchwKcyxNkhw; rename static ck source files git-subtree-dir: src/composable_kernel git-subtree-split: 31b403526ec54abf13c4bb58dfb6635b4d2aa619
ROCm · Aug 21, 2021 · 3952bef · 3952bef
1 parent 437cc59
commit 3952bef
Show file tree

Hide file tree

Showing 151 changed files with 5,790 additions and 16,965 deletions.
diff --git a/.clang-tidy b/.clang-tidy
@@ -0,0 +1,3 @@
+CheckOptions:
+  - key: bugprone-reserved-identifier.AllowedIdentifiers
+    value: '__HIP_PLATFORM_HCC__;__HIP_ROCclr__'
diff --git a/CMakeLists.txt b/CMakeLists.txt
@@ -1,10 +1,9 @@
-cmake_minimum_required(VERSION 2.8.3)
-project(modular_convolution)
+cmake_minimum_required(VERSION 3.5)
+project(composable_kernel)
 
 list(APPEND CMAKE_MODULE_PATH "${PROJECT_SOURCE_DIR}/cmake")
 
-include(TargetFlags)
-include(AddKernels)
+include(CheckCXXCompilerFlag)
 
 ## C++
 enable_language(CXX)
@@ -39,4 +38,161 @@ link_libraries(${OpenMP_pthread_LIBRARY})
 find_package(HIP REQUIRED)
 message(STATUS "Build with HIP ${hip_VERSION}")
 
+## half
+#find_path(HALF_INCLUDE_DIR half.hpp)
+message("HALF_INCLUDE_DIR: ${HALF_INCLUDE_DIR}")
+
+# CMAKE_CXX_FLAGS
+SET(BUILD_DEV ON CACHE BOOL "BUILD_DEV")
+if(BUILD_DEV)
+    string(APPEND CMAKE_CXX_FLAGS " -Werror -Weverything")
+endif()
+message("CMAKE_CXX_FLAGS: ${CMAKE_CXX_FLAGS}")
+
+## tidy
+include(EnableCompilerWarnings)
+set(MIOPEN_TIDY_ERRORS ERRORS * -readability-inconsistent-declaration-parameter-name)
+if(CMAKE_CXX_COMPILER MATCHES ".*hcc" OR CMAKE_CXX_COMPILER MATCHES ".*clang\\+\\+")
+    set(MIOPEN_TIDY_CHECKS -modernize-use-override -readability-non-const-parameter)
+# Enable tidy on hip
+elseif(MIOPEN_BACKEND STREQUAL "HIP" OR MIOPEN_BACKEND STREQUAL "HIPNOGPU")
+    set(MIOPEN_TIDY_ERRORS ALL)
+endif()
+
+include(ClangTidy)
+enable_clang_tidy(
+    CHECKS
+        *
+        -abseil-*
+        -android-cloexec-fopen
+        # Yea we shouldn't be using rand()
+        -cert-msc30-c
+        -bugprone-exception-escape
+        -bugprone-macro-parentheses
+        -cert-env33-c
+        -cert-msc32-c
+        -cert-msc50-cpp
+        -cert-msc51-cpp
+        -cert-dcl37-c
+        -cert-dcl51-cpp
+        -clang-analyzer-alpha.core.CastToStruct
+        -clang-analyzer-optin.performance.Padding
+        -clang-diagnostic-deprecated-declarations
+        -clang-diagnostic-extern-c-compat
+        -clang-diagnostic-unused-command-line-argument
+        -cppcoreguidelines-avoid-c-arrays
+        -cppcoreguidelines-avoid-magic-numbers
+        -cppcoreguidelines-explicit-virtual-functions
+        -cppcoreguidelines-init-variables
+        -cppcoreguidelines-macro-usage
+        -cppcoreguidelines-non-private-member-variables-in-classes
+        -cppcoreguidelines-pro-bounds-array-to-pointer-decay
+        -cppcoreguidelines-pro-bounds-constant-array-index
+        -cppcoreguidelines-pro-bounds-pointer-arithmetic
+        -cppcoreguidelines-pro-type-member-init
+        -cppcoreguidelines-pro-type-reinterpret-cast
+        -cppcoreguidelines-pro-type-union-access
+        -cppcoreguidelines-pro-type-vararg
+        -cppcoreguidelines-special-member-functions
+        -fuchsia-*
+        -google-explicit-constructor
+        -google-readability-braces-around-statements
+        -google-readability-todo
+        -google-runtime-int
+        -google-runtime-references
+        -hicpp-vararg
+        -hicpp-braces-around-statements
+        -hicpp-explicit-conversions
+        -hicpp-named-parameter
+        -hicpp-no-array-decay
+        # We really shouldn't use bitwise operators with signed integers, but
+        # opencl leaves us no choice
+        -hicpp-avoid-c-arrays
+        -hicpp-signed-bitwise
+        -hicpp-special-member-functions
+        -hicpp-uppercase-literal-suffix
+        -hicpp-use-auto
+        -hicpp-use-equals-default
+        -hicpp-use-override
+        -llvm-header-guard
+        -llvm-include-order
+        #-llvmlibc-*
+        -llvmlibc-restrict-system-libc-headers
+        -llvmlibc-callee-namespace
+        -llvmlibc-implementation-in-namespace
+        -llvm-else-after-return
+        -llvm-qualified-auto
+        -misc-misplaced-const
+        -misc-non-private-member-variables-in-classes
+        -misc-no-recursion
+        -modernize-avoid-bind
+        -modernize-avoid-c-arrays
+        -modernize-pass-by-value
+        -modernize-use-auto
+        -modernize-use-default-member-init
+        -modernize-use-equals-default
+        -modernize-use-trailing-return-type
+        -modernize-use-transparent-functors
+        -performance-unnecessary-value-param
+        -readability-braces-around-statements
+        -readability-else-after-return
+        # we are not ready to use it, but very useful
+        -readability-function-cognitive-complexity
+        -readability-isolate-declaration
+        -readability-magic-numbers
+        -readability-named-parameter
+        -readability-uppercase-literal-suffix
+        -readability-convert-member-functions-to-static
+        -readability-qualified-auto
+        -readability-redundant-string-init
+        # too many narrowing conversions in our code
+        -bugprone-narrowing-conversions
+        -cppcoreguidelines-narrowing-conversions
+        -altera-struct-pack-align
+        -cppcoreguidelines-prefer-member-initializer
+
+        ${MIOPEN_TIDY_CHECKS}
+        ${MIOPEN_TIDY_ERRORS}
+    HEADER_FILTER
+        "\.hpp$"
+    EXTRA_ARGS
+        -DMIOPEN_USE_CLANG_TIDY
+)
+
+include(CppCheck)
+enable_cppcheck(
+    CHECKS
+        warning
+        style
+        performance
+        portability
+    SUPPRESS
+        ConfigurationNotChecked
+        constStatement
+        duplicateCondition
+        noExplicitConstructor
+        passedByValue
+        preprocessorErrorDirective
+        shadowVariable
+        unusedFunction
+        unusedPrivateFunction
+        unusedStructMember
+        unmatchedSuppression
+    FORCE
+    SOURCES
+        host/host_tensor/src
+        host/driver_offline/src
+        composable_kernel/src/kernel_wrapper
+    INCLUDE
+        host/host_tensor/include
+        host/solver/include
+        host/driver_offline/include
+        composable_kernel/include/*
+        ${CMAKE_CURRENT_SOURCE_DIR}/include
+        ${CMAKE_CURRENT_BINARY_DIR}/include
+    DEFINE
+        CPPCHECK=1
+        __linux__=1
+)
+
 add_subdirectory(host)
diff --git a/README.md b/README.md
@@ -78,7 +78,7 @@ InLeftPads size 2, {1, 1, }
 InRightPads size 2, {1, 1, }
 ConvStrides size 2, {2, 2, }
 ConvDilations size 2, {1, 1, }
-device_dynamic_convolution_forward_implicit_gemm_v4r4r2_xdlops_nchw_kcyx_nkhw
+device_convolution_forward_implicit_gemm_v4r4r2_xdlops_nchw_kcyx_nkhw
 a_k0_m_k1_grid_desc{216, 256, 8}
 b_k0_n_k1_grid_desc{216, 165888, 8}
 c_m_n_grid_desc{ 256, 165888}
@@ -100,7 +100,7 @@ InLeftPads size 2, {1, 1, }
 InRightPads size 2, {1, 1, }
 ConvStrides size 2, {1, 1, }
 ConvDilations size 2, {1, 1, }
-device_dynamic_convolution_forward_implicit_gemm_v4r4r2_xdlops_nchw_kcyx_nkhw
+device_convolution_forward_implicit_gemm_v4r4r2_xdlops_nchw_kcyx_nkhw
 a_k0_m_k1_grid_desc{288, 1024, 8}
 b_k0_n_k1_grid_desc{288, 50176, 8}
 c_m_n_grid_desc{ 1024, 50176}
@@ -122,7 +122,7 @@ InLeftPads size 2, {1, 1, }
 InRightPads size 2, {1, 1, }
 ConvStrides size 2, {2, 2, }
 ConvDilations size 2, {1, 1, }
-device_dynamic_convolution_forward_implicit_gemm_v4r4r4_xdlops_nhwc_kyxc_nhwk
+device_convolution_forward_implicit_gemm_v4r4r4_xdlops_nhwc_kyxc_nhwk
 a_k0_m_k1_grid_desc{216, 165888, 8}
 b_k0_n_k1_grid_desc{216, 256, 8}
 c_m_n_grid_desc{ 165888, 256}
@@ -144,7 +144,7 @@ InLeftPads size 2, {1, 1, }
 InRightPads size 2, {1, 1, }
 ConvStrides size 2, {1, 1, }
 ConvDilations size 2, {1, 1, }
-device_dynamic_convolution_forward_implicit_gemm_v4r4r4_xdlops_nhwc_kyxc_nhwk
+device_convolution_forward_implicit_gemm_v4r4r4_xdlops_nhwc_kyxc_nhwk
 a_k0_m_k1_grid_desc{288, 50176, 8}
 b_k0_n_k1_grid_desc{288, 1024, 8}
 c_m_n_grid_desc{ 50176, 1024}
@@ -166,7 +166,7 @@ InLeftPads size 2, {1, 1, }
 InRightPads size 2, {1, 1, }
 ConvStrides size 2, {1, 1, }
 ConvDilations size 2, {1, 1, }
-device_dynamic_convolution_backward_data_implicit_gemm_v4r1r2_xdlops_nhwc_kyxc_nhwk
+device_convolution_backward_data_implicit_gemm_v4r1r2_xdlops_nhwc_kyxc_nhwk
 a_k0_m_k1_grid_desc{288, 50176, 8}
 b_k0_n_k1_grid_desc{288, 1024, 8}
 c_m_n_grid_desc{ 50176, 1024}

diff --git a/cmake/AddKernels.cmake b/cmake/AddKernels.cmake
diff --git a/.../online_compile/addkernels/CMakeLists.txt → cmake/Analyzers.cmake b/.../online_compile/addkernels/CMakeLists.txt → cmake/Analyzers.cmake
@@ -24,7 +24,11 @@
 # 
 ################################################################################
 
-set(ADD_KERNELS_SOURCE include_inliner.cpp addkernels.cpp)
+if(NOT TARGET analyze)
+    add_custom_target(analyze)
+endif()
 
-add_executable(addkernels EXCLUDE_FROM_ALL ${ADD_KERNELS_SOURCE})
+function(mark_as_analyzer)
+    add_dependencies(analyze ${ARGN})
+endfunction()