Update parsing of CMake flags to handle generator expressions (#1521)

* cmake-flag-parsing-update(01) Apply changes from "Improve handling of generator expressions when getting the flags for hip", ROCm/AMDMIGraphX#1055
ROCm · Apr 15, 2022 · a8349c0 · a8349c0
1 parent 86f12f4
commit a8349c0
Show file tree

Hide file tree

Showing 3 changed files with 56 additions and 66 deletions.
diff --git 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 "$<LINK_LANGUAGE:CXX>" "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
@@ -1,38 +1,73 @@
 
+function(eval_and_strip_genex OUTPUT_VAR INPUT)
+    string(REPLACE "$<LINK_LANGUAGE:CXX>" "1" INPUT "${INPUT}")
+    string(REPLACE "$<COMPILE_LANGUAGE:CXX>" "1" INPUT "${INPUT}")
+    string(REPLACE "SHELL:" "" INPUT "${INPUT}")
+    string(REPLACE "$<BOOL:>" "0" INPUT "${INPUT}")
+    string(REGEX REPLACE "\\$<BOOL:(0|FALSE|false|OFF|off|N|n|IGNORE|ignore|NOTFOUND|notfound)>" "0" INPUT "${INPUT}")
+    string(REGEX REPLACE "\\$<BOOL:[^<>]*-NOTFOUND>" "0" INPUT "${INPUT}")
+    string(REGEX REPLACE "\\$<BOOL:[^$<>]*>" "1" INPUT "${INPUT}")
+    string(REPLACE "$<NOT:0>" "1" INPUT "${INPUT}")
+    string(REPLACE "$<NOT:1>" "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()

diff --git 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<TmpDir>& tmp_dir,
                                             const std::string& filename,
                                             std::string src,
@@ -125,14 +83,8 @@ static boost::filesystem::path HipBuildImpl(boost::optional<TmpDir>& 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";