From 5e9566d82afbf9e11a4cec4f9b9fd2ba706f0e53 Mon Sep 17 00:00:00 2001 From: Artem Tamazov Date: Thu, 14 Apr 2022 21:41:34 +0300 Subject: [PATCH 1/2] cmake-flag-parsing-update(01) Apply changes from "Improve handling of generator expressions when getting the flags for hip", https://github.com/ROCmSoftwarePlatform/AMDMIGraphX/pull/1055 --- CMakeLists.txt | 9 ++++-- cmake/TargetFlags.cmake | 63 ++++++++++++++++++++++++++++++++--------- 2 files changed, 55 insertions(+), 17 deletions(-) diff --git a/CMakeLists.txt b/CMakeLists.txt index 18a059ace9..2bc54cbe85 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -192,9 +192,12 @@ message(STATUS "Build with HIP ${hip_VERSION}") target_flags(HIP_COMPILER_FLAGS hip::device) # Remove cuda arch flags string(REGEX REPLACE --cuda-gpu-arch=[a-z0-9]+ "" HIP_COMPILER_FLAGS "${HIP_COMPILER_FLAGS}") -string(REGEX REPLACE --offload-arch=[a-z0-9]+ "" HIP_COMPILER_FLAGS "${HIP_COMPILER_FLAGS}") -string(REPLACE "$" "1" HIP_COMPILER_FLAGS "${HIP_COMPILER_FLAGS}") -string(REPLACE "SHELL:" "" HIP_COMPILER_FLAGS "${HIP_COMPILER_FLAGS}") +string(REGEX REPLACE --offload-arch=[a-z0-9:+-]+ "" HIP_COMPILER_FLAGS "${HIP_COMPILER_FLAGS}") +# Skip library paths since hip will incorrectly treat it as a source file +string(APPEND HIP_COMPILER_FLAGS " ") +foreach(_unused RANGE 2) + string(REGEX REPLACE " /[^ ]+\\.(a|so) " " " HIP_COMPILER_FLAGS "${HIP_COMPILER_FLAGS}") +endforeach() # Override HIP version in config.h, if necessary. # The variables set by find_package() can't be overwritten, diff --git a/cmake/TargetFlags.cmake b/cmake/TargetFlags.cmake index 4f83fb5d39..ede876fb55 100644 --- a/cmake/TargetFlags.cmake +++ b/cmake/TargetFlags.cmake @@ -1,38 +1,73 @@ +function(eval_and_strip_genex OUTPUT_VAR INPUT) + string(REPLACE "$" "1" INPUT "${INPUT}") + string(REPLACE "$" "1" INPUT "${INPUT}") + string(REPLACE "SHELL:" "" INPUT "${INPUT}") + string(REPLACE "$" "0" INPUT "${INPUT}") + string(REGEX REPLACE "\\$" "0" INPUT "${INPUT}") + string(REGEX REPLACE "\\$]*-NOTFOUND>" "0" INPUT "${INPUT}") + string(REGEX REPLACE "\\$]*>" "1" INPUT "${INPUT}") + string(REPLACE "$" "1" INPUT "${INPUT}") + string(REPLACE "$" "0" INPUT "${INPUT}") + string(REGEX REPLACE "\\$<0:[^<>]*>" "" INPUT "${INPUT}") + string(REGEX REPLACE "\\$<1:([^<>]*)>" "\\1" INPUT "${INPUT}") + string(GENEX_STRIP "${INPUT}" INPUT) + set(${OUTPUT_VAR} "${INPUT}" PARENT_SCOPE) +endfunction() + function(get_target_property2 VAR TARGET PROPERTY) get_target_property(_pflags ${TARGET} ${PROPERTY}) if(_pflags) + eval_and_strip_genex(_pflags "${_pflags}") set(${VAR} ${_pflags} PARENT_SCOPE) else() set(${VAR} "" PARENT_SCOPE) endif() endfunction() +function(flags_requires_arg OUTPUT_VAR FLAG) + set(_args -x -isystem) + if(FLAG IN_LIST _args) + set(${OUTPUT_VAR} 1 PARENT_SCOPE) + else() + set(${OUTPUT_VAR} 0 PARENT_SCOPE) + endif() +endfunction() macro(append_flags FLAGS TARGET PROPERTY PREFIX) get_target_property2(_pflags ${TARGET} ${PROPERTY}) + set(_requires_arg 0) foreach(FLAG ${_pflags}) - if(TARGET ${FLAG}) - target_flags(_pflags2 ${FLAG}) - string(APPEND ${FLAGS} " ${_pflags2}") - else() - string(APPEND ${FLAGS} " ${PREFIX}${FLAG}") + string(STRIP "${FLAG}" FLAG) + if(FLAG) + if(TARGET ${FLAG} AND NOT _requires_arg) + target_flags(_pflags2 ${FLAG}) + string(APPEND ${FLAGS} " ${_pflags2}") + else() + string(APPEND ${FLAGS} " ${PREFIX}${FLAG}") + endif() + flags_requires_arg(_requires_arg "${FLAG}") endif() endforeach() endmacro() macro(append_link_flags FLAGS TARGET PROPERTY) get_target_property2(_pflags ${TARGET} ${PROPERTY}) + set(_requires_arg 0) foreach(FLAG ${_pflags}) - if(TARGET ${FLAG}) - target_flags(_pflags2 ${FLAG}) - string(APPEND ${FLAGS} " ${_pflags2}") - elseif(FLAG MATCHES "^-.*") - string(APPEND ${FLAGS} " ${FLAG}") - elseif(EXISTS ${FLAG}) - string(APPEND ${FLAGS} " ${FLAG}") - else() - string(APPEND ${FLAGS} " -l${FLAG}") + string(STRIP "${FLAG}" FLAG) + if(FLAG) + if(TARGET ${FLAG} AND NOT _requires_arg) + target_flags(_pflags2 ${FLAG}) + string(APPEND ${FLAGS} " ${_pflags2}") + elseif(FLAG MATCHES "^-.*") + string(APPEND ${FLAGS} " ${FLAG}") + elseif(EXISTS ${FLAG}) + string(APPEND ${FLAGS} " ${FLAG}") + else() + string(APPEND ${FLAGS} " -l${FLAG}") + endif() + flags_requires_arg(_requires_arg "${FLAG}") endif() endforeach() endmacro() From 5f1eda0b1646a643ab147df546fa5806b4eb1b41 Mon Sep 17 00:00:00 2001 From: Artem Tamazov Date: Thu, 14 Apr 2022 21:51:37 +0300 Subject: [PATCH 2/2] cmake-flag-parsing-update(02) Remove some code which is useless now --- src/hip/hip_build_utils.cpp | 50 +------------------------------------ 1 file changed, 1 insertion(+), 49 deletions(-) diff --git a/src/hip/hip_build_utils.cpp b/src/hip/hip_build_utils.cpp index 7c3ede275c..4f38d33f50 100644 --- a/src/hip/hip_build_utils.cpp +++ b/src/hip/hip_build_utils.cpp @@ -42,48 +42,6 @@ MIOPEN_DECLARE_ENV_VAR(MIOPEN_DEBUG_HIP_DUMP) namespace miopen { -namespace { - -inline void InplaceRemoveAllTokensThatInclude(std::string& str, - const std::string& search_1, - const std::string& search_2) -{ - for(std::size_t pos = 0;;) - { - if((pos = str.find(search_1, pos)) == std::string::npos) - break; - - auto beg = str.find_last_of(" \t\n\r", pos); - beg = (beg == std::string::npos) ? 0 : beg + 1; - const auto end = str.find_first_of(" \t\n\r", pos); - - auto len = std::string::npos; - auto next = std::string::npos; - if(end != std::string::npos) - { - next = str.find_first_not_of(" \t\n\r", end); - if(next != std::string::npos) - len = next - beg; - } - - const auto tok = - end == std::string::npos ? std::string{str, beg} : std::string{str, beg, end - beg}; - if(tok.find(search_2) != std::string::npos) - { - str.erase(beg, len); - if(len == std::string::npos) - break; - pos = beg; - } - else - { - pos = next; - } - } -} - -} // namespace - static boost::filesystem::path HipBuildImpl(boost::optional& tmp_dir, const std::string& filename, std::string src, @@ -125,14 +83,8 @@ static boost::filesystem::path HipBuildImpl(boost::optional& tmp_dir, params += " --cuda-device-only"; params += " -c"; params += " -O3 "; - params += " -Wno-unused-command-line-argument -I. "; - - { - std::string hip_compiler_flags = MIOPEN_STRINGIZE(HIP_COMPILER_FLAGS); - InplaceRemoveAllTokensThatInclude(hip_compiler_flags, "clang_rt.builtins", "x86_64"); - params += hip_compiler_flags; - } + params += MIOPEN_STRINGIZE(HIP_COMPILER_FLAGS); #if HIP_PACKAGE_VERSION_FLAT < 4004000000ULL params += " -mllvm --amdgpu-spill-vgpr-to-agpr=0";