Improved SYCL flag handling

The use of gmx_check_source_compiles_with_flags has been improved to provide more granularity in the case of failure. COMPILE_FLAGS source property has been superseded by COMPILE_OPTIONS, which also works better. For example, we no longer need to strip strings of flags. target_link_libraries() is no longer being abused to pass flags, which will make it possible to use link dependencies with module object libraries that include SYCL code. Compile-time only flags are no longer used at link time. Improved docs of gmx_check_source_compiles_with_flags() and gmx_check_source_compiles_with_flags().
kassonlab · Jul 29, 2022 · dea311a · dea311a
1 parent 9c87c79
commit dea311a
Show file tree

Hide file tree

Showing 2 changed files with 59 additions and 20 deletions.
diff --git a/cmake/gmxFindFlagsForSource.cmake b/cmake/gmxFindFlagsForSource.cmake
@@ -44,7 +44,9 @@ function(gmx_check_compiler_flag FLAGS LANGUAGE RESULT_VARIABLE)
 endfunction()
 
 # Helper function to call the correct language version of the CMake
-# check_*_source_compiles function.
+# check_*_source_compiles function. The use of a function creates a
+# scope that prevents the leaking of the flags via
+# CMAKE_REQUIRED_FLAGS.
 function(gmx_check_source_compiles_with_flags SOURCE FLAGS LANGUAGE RESULT_VARIABLE)
    set(CMAKE_REQUIRED_FLAGS "${FLAGS}")
    if (LANGUAGE STREQUAL "C")
@@ -56,7 +58,15 @@ function(gmx_check_source_compiles_with_flags SOURCE FLAGS LANGUAGE RESULT_VARIA
    endif()
 endfunction()
 
-# Helper routine to find flag (from a list) to compile a specific source (in C or C++).
+# Helper routine to find flag (from a list) to compile a specific
+# source (in C or C++).  It loops over flags in the list, first
+# checking that the compiler recognizes the flag, and then that it can
+# compile the given sources with that flag. When it finds a flag, it
+# leaves the loop, otherwise it proceeds to the next flag in the list.
+#
+# If you want to end up passing multiple flags, call this or similar
+# methods once for each flag.
+#
 # RESULT_VARIABLE           Name of variable to set in the parent scope to true if
 #                           we have found a flag that works (which could be "")
 # SOURCE                    Source code to test

diff --git a/cmake/gmxManageSYCL.cmake b/cmake/gmxManageSYCL.cmake
@@ -264,7 +264,7 @@ else()
             DISABLE_SYCL_CXX_FLAGS_RESULT)
 
         if(DISABLE_SYCL_CXX_FLAGS_RESULT)
-            set(DISABLE_SYCL_CXX_FLAGS "-fno-sycl")
+            set(SYCL_TOOLCHAIN_CXX_FLAGS "-fno-sycl")
         endif()
         if(NOT CHECK_DISABLE_SYCL_CXX_FLAGS_QUIETLY)
             if(DISABLE_SYCL_CXX_FLAGS_RESULT)
@@ -275,34 +275,59 @@ else()
             set(CHECK_DISABLE_SYCL_CXX_FLAGS_QUIETLY 1 CACHE INTERNAL "Keep quiet on future calls to detect no-SYCL flags" FORCE)
         endif()
     endif()
-    
+
     # Find the flags to enable (or re-enable) SYCL with Intel extensions. In case we turned it off above,
     # it's important that we check the combination of both flags, to make sure the second one re-enables SYCL.
     if(NOT CHECK_SYCL_CXX_FLAGS_QUIETLY)
         message(STATUS "Checking for flags to enable SYCL")
     endif()
-    gmx_find_flag_for_source(SYCL_CXX_FLAGS_RESULT
+    set(SAMPLE_SYCL_SOURCE
         "#include <CL/sycl.hpp>
          int main(){
              sycl::queue q(sycl::default_selector{});
              return 0;
-         }
-         " "CXX" DISABLE_SYCL_CXX_FLAGS SYCL_CXX_FLAGS " -fsycl -fsycl-device-code-split=per_kernel ${SYCL_CXX_FLAGS_EXTRA}")
-
-    string(STRIP "${SYCL_CXX_FLAGS}" SYCL_CXX_FLAGS)
-    if(NOT CHECK_SYCL_CXX_FLAGS_QUIETLY)
-        if(SYCL_CXX_FLAGS_RESULT)
-            message(STATUS "Checking for flags to enable SYCL - ${SYCL_CXX_FLAGS}")
+         }")
+    set(SYCL_CXX_FLAGS "-fsycl")
+    gmx_check_source_compiles_with_flags(
+        "${SAMPLE_SYCL_SOURCE}"
+        "${SYCL_TOOLCHAIN_CXX_FLAGS} ${SYCL_CXX_FLAGS}"
+        "CXX"
+        SYCL_CXX_FLAGS_RESULT
+        )
+    if (SYCL_CXX_FLAGS_RESULT)
+        if(NOT CHECK_SYCL_CXX_FLAGS_QUIETLY)
+            message(STATUS "Checking for flags to enable SYCL - ${SYCL_TOOLCHAIN_CXX_FLAGS} ${SYCL_CXX_FLAGS}")
         endif()
         set(CHECK_SYCL_CXX_FLAGS_QUIETLY 1 CACHE INTERNAL "Keep quiet on future calls to detect SYCL flags" FORCE)
+        set(SYCL_TOOLCHAIN_CXX_FLAGS "${SYCL_TOOLCHAIN_CXX_FLAGS} ${SYCL_CXX_FLAGS}")
+    else()
+        message(FATAL_ERROR "Cannot compile a SYCL program with ${SYCL_TOOLCHAIN_CXX_FLAGS} ${SYCL_CXX_FLAGS}. Try a different compiler or disable SYCL.")
     endif()
-
-    if(NOT SYCL_CXX_FLAGS_RESULT)
-        message(FATAL_ERROR "Cannot compile with SYCL Intel compiler. Try a different compiler or disable SYCL.")
+
+    # Add kernel-splitting flag if available
+    set(SYCL_DEVICE_CODE_SPLIT_CXX_FLAGS "-fsycl-device-code-split=per_kernel")
+    gmx_check_source_compiles_with_flags(
+        "${SAMPLE_SYCL_SOURCE}"
+        "${SYCL_TOOLCHAIN_CXX_FLAGS} ${SYCL_DEVICE_CODE_SPLIT_CXX_FLAGS}"
+        "CXX"
+        SYCL_DEVICE_CODE_SPLIT_CXX_FLAGS_RESULT
+        )
+    if (SYCL_DEVICE_CODE_SPLIT_CXX_FLAGS_RESULT)
+        set(SYCL_TOOLCHAIN_CXX_FLAGS "${SYCL_TOOLCHAIN_CXX_FLAGS} ${SYCL_DEVICE_CODE_SPLIT_CXX_FLAGS}")
+    else()
+        message(WARNING "Cannot compile SYCL with per-kernel device-code splitting. Simulations will work, but the first step will be much slower than it needs to be. Try a different compiler.")
     endif()
 
-    if(NOT WIN32)
-         set(SYCL_CXX_FLAGS "${SYCL_CXX_FLAGS} -ffast-math")
+    # Add fast-math flag where available
+    gmx_find_flag_for_source(
+        SYCL_FAST_MATH_CXX_FLAGS_RESULT
+        "${SAMPLE_SYCL_SOURCE}"
+        "CXX"
+        SYCL_TOOLCHAIN_CXX_FLAGS
+        SYCL_FAST_MATH_CXX_FLAGS
+        "-ffast-math" "/clang:-ffast-math")
+    if (SYCL_FAST_MATH_CXX_FLAGS_RESULT)
+        set(SYCL_TOOLCHAIN_CXX_FLAGS "${SYCL_TOOLCHAIN_CXX_FLAGS} ${SYCL_FAST_MATH_CXX_FLAGS}")
     endif()
 
     include(gmxManageFFTLibraries)
@@ -321,8 +346,12 @@ else()
             "" # No options
             "TARGET" # One-value keyword
             "SOURCES" # Multi-value keyword
-        )
-        set_source_files_properties(${ARGS_SOURCES} PROPERTIES COMPILE_FLAGS "${SYCL_CXX_FLAGS}")
-        target_link_libraries(${ARGS_TARGET} PRIVATE ${SYCL_CXX_FLAGS})
+            )
+        # convert the space-separated string to a list
+        separate_arguments(SYCL_TOOLCHAIN_CXX_FLAGS)
+        set_property(SOURCE ${ARGS_SOURCES} APPEND PROPERTY COMPILE_OPTIONS
+            ${SYCL_TOOLCHAIN_CXX_FLAGS}
+            ${SYCL_CXX_FLAGS_EXTRA})
+        target_link_options(${ARGS_TARGET} PRIVATE ${SYCL_CXX_FLAGS})
     endfunction(add_sycl_to_target)
 endif()