[adaptivecpp] Add support for SYCL with adaptiveCpp

Author: devajithvs

.github/workflows/root-ci-config/buildconfig/global.txt

@@ -1,3 +1,4 @@
+adaptivecpp=ON
 alien=OFF
 all=OFF
 arrow=OFF

cmake/modules/RootBuildOptions.cmake

@@ -77,6 +77,7 @@ endfunction()
 #   The default value can be changed as many times as we wish before calling ROOT_APPLY_OPTIONS()
 #--------------------------------------------------------------------------------------------------
 
+ROOT_BUILD_OPTION(adaptivecpp ON "Build AdaptiveCPP, for SYCL support")
 ROOT_BUILD_OPTION(arrow OFF "Enable support for Apache Arrow")
 ROOT_BUILD_OPTION(asimage ON "Enable support for image processing via libAfterImage")
 ROOT_BUILD_OPTION(asserts OFF "Enable asserts (defaults to ON for CMAKE_BUILD_TYPE=Debug and/or dev=ON)")

cmake/modules/SearchInstalledSoftware.cmake

@@ -61,6 +61,11 @@ macro(ROOT_CHECK_CONNECTION_AND_DISABLE_OPTION option_name)
   endif()
 endmacro()
 
+# Building adaptivecpp requires an internet connection, if we're not side-loading the source directory
+if(adaptivecpp AND NOT DEFINED ADAPTIVECPP_SOURCE_DIR)
+  ROOT_CHECK_CONNECTION_AND_DISABLE_OPTION("adaptivecpp")
+endif()
+
 # Building Clad requires an internet connection, if we're not side-loading the source directory
 if(clad AND NOT DEFINED CLAD_SOURCE_DIR)
   ROOT_CHECK_CONNECTION_AND_DISABLE_OPTION("clad")

core/metacling/src/CMakeLists.txt

@@ -108,7 +108,8 @@ if (clad)
   else()
     set(CLING_PLUGIN_LINK_LIBS -Wl,--whole-archive cladPlugin cladDifferentiator -Wl,--no-whole-archive)
   endif()
-  if(TARGET clang)
+  # This will cause a cyclic dependency in adaptiveCpp, as it requires clang to build
+  if(TARGET clang AND NOT adaptivecpp)
     # Link our clad libraries to clang. If users use the clang from ROOT they will
     # also be able to use clad out of the box.
     add_dependencies(clang clad)

interpreter/CMakeLists.txt

@@ -53,7 +53,10 @@ set(LLVM_FORCE_USE_OLD_TOOLCHAIN ON CACHE BOOL "")
 # found before (because the user turned on cuda or tmva-gpu), this will not have
 # an effect, which is fine.
 # (Note that the option is very counter-intuitive: We turn *on* disabling it...)
-set(CMAKE_DISABLE_FIND_PACKAGE_CUDA ON)
+# Special case when we want to build cling with adaptiveCpp
+if(NOT adaptivecpp)
+  set(CMAKE_DISABLE_FIND_PACKAGE_CUDA ON)
+endif()
 
 # will be set again in case NOT builtin_llvm
 set(LLVM_DIR "${CMAKE_CURRENT_BINARY_DIR}/llvm-project/llvm")
@@ -120,7 +123,7 @@ if(clingtest)
   endif()
   ROOT_ADD_TEST(clingtest-check-cling COMMAND ${CMAKE_COMMAND} --build ${CMAKE_BINARY_DIR} --target check-cling
                                        ENVIRONMENT ${CLINGTEST_EXECUTABLE})
-else()
+elseif(NOT adaptivecpp) # AdaptiveCPP needs hipsycl::rt symbols to be exposed
   #---Build LLVM/Clang with symbol visibility=hidden--------------------------------------------------
   set(CMAKE_CXX_VISIBILITY_PRESET hidden)
   set(CMAKE_C_VISIBILITY_PRESET hidden)
@@ -390,6 +393,9 @@ if (builtin_clang)
     ${CLANG_BINARY_DIR}/include
     CACHE STRING "Clang include directories.")
   set(CLANG_CMAKE_DIR "${CMAKE_BINARY_DIR}/lib/cmake/clang/")
+  set(CLANG_EXECUTABLE_PATH
+    "${LLVM_BINARY_DIR}/bin/clang${CMAKE_EXECUTABLE_SUFFIX}"
+    CACHE STRING "Path to just‐built clang (if builtin_clang).")
 else()
   find_package(Clang REQUIRED CONFIG)
   message(STATUS "Found Clang ${CLANG_PACKAGE_VERSION} in ${CLANG_CMAKE_DIR}")

interpreter/cling/CMakeLists.txt

@@ -128,6 +128,21 @@ set(CLING_BINARY_DIR ${CMAKE_CURRENT_BINARY_DIR})
 set(LLVM_TOOLS_BINARY_DIR "${LLVM_BINARY_DIR}/bin")
 if(DEFINED CLING_ROOT_BUILD)
   set(CMAKE_RUNTIME_OUTPUT_DIRECTORY "${LLVM_TOOLS_BINARY_DIR}")
+
+  if(adaptivecpp)
+    # Hack for now to make it work with adaptiveCpp without making more changes
+    # Ensure runtime ACPP directory exists to prevent SYCL from running
+    set(ACPP_RUNTIME_DIRECTORY "${CMAKE_RUNTIME_OUTPUT_DIRECTORY}/../lib/hipSYCL/")
+    file(MAKE_DIRECTORY ${ACPP_RUNTIME_DIRECTORY})
+    file(MAKE_DIRECTORY ${CMAKE_BINARY_DIR}/lib/hipSYCL/bitcode)
+
+    # Create a symbolic link to the bitcode directory
+    execute_process(
+      COMMAND ${CMAKE_COMMAND} -E create_symlink
+      "${CMAKE_BINARY_DIR}/lib/hipSYCL/bitcode"
+      "${ACPP_RUNTIME_DIRECTORY}/bitcode"
+    )
+  endif()
 endif()
 
 if( CMAKE_SOURCE_DIR STREQUAL CMAKE_BINARY_DIR AND NOT MSVC_IDE )
@@ -236,7 +251,7 @@ if(MSVC)
       set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -std:c++14")
     endif()
   endif()
-else()
+elseif(NOT adaptivecpp) # adaptivecpp does not build with -fno-rtti
   set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -fno-rtti")
 endif()
 
@@ -470,6 +485,60 @@ if (TARGET clang-headers)
   list(APPEND LLVM_COMMON_DEPENDS clang-headers)
 endif()
 
+if(adaptivecpp)
+  # Use ADAPTIVE_CPP_SOURCE_DIR variable if it is defined in the CMake
+  # configuration
+  if(NOT DEFINED ADAPTIVE_CPP_SOURCE_DIR)
+    set(ADAPTIVE_CPP_SOURCE_DIR ${CMAKE_CURRENT_SOURCE_DIR}/AdaptiveCpp)
+  endif()
+
+  set(HIPSYCL_NO_FIBERS ON)
+  set(WITH_OPENCL_BACKEND OFF)
+  set(WITH_LEVEL_ZERO_BACKEND OFF)
+
+  add_compile_definitions(CLING_WITH_ADAPTIVECPP)
+
+  list(INSERT CMAKE_MODULE_PATH 0 "${ADAPTIVE_CPP_SOURCE_DIR}/cmake/")
+
+  set(ADAPTIVE_CPP_HEADER_BUILD_PATH
+      "${CMAKE_CURRENT_BINARY_DIR}/include/AdaptiveCpp")
+  set(ADAPTIVE_CPP_HEADER_INSTALL_PATH
+      "${CMAKE_INSTALL_PREFIX}/include/AdaptiveCpp")
+
+  add_subdirectory(${ADAPTIVE_CPP_SOURCE_DIR})
+  set(ACPP_EXPORTS
+      acpp-common
+      acpp-rt
+      acpp-clang-cbs
+      acpp-clang
+      llvm-to-backend
+      llvm-to-host
+      rt-backend-omp)
+
+  if(WITH_CUDA_BACKEND)
+    list(APPEND ACPP_EXPORTS llvm-to-ptx rt-backend-cuda)
+  endif()
+
+  if(WITH_OPENCL_BACKEND)
+    list(APPEND ACPP_EXPORTS llvm-to-spirv rt-backend-ocl ocl-cxx-headers ocl-headers)
+  endif()
+
+  if(WITH_ROCM_BACKEND)
+    list(APPEND ACPP_EXPORTS llvm-to-amdgpu rt-backend-hip)
+  endif()
+
+  foreach(target IN LISTS ACPP_EXPORTS)
+    install(TARGETS ${target} EXPORT ClingTargets)
+  endforeach()
+
+  set_property(GLOBAL APPEND PROPERTY CLING_EXPORTS ${ACPP_EXPORTS})
+
+  if(TARGET intrinsics_gen)
+    add_dependencies(acpp-clang-cbs intrinsics_gen)
+    add_dependencies(llvm-to-backend intrinsics_gen)
+  endif()
+endif()
+
 add_subdirectory(include/cling/Interpreter)
 add_subdirectory(lib)
 

interpreter/cling/lib/Interpreter/BackendPasses.cpp

@@ -29,6 +29,12 @@
 #include "llvm/Transforms/Scalar.h"
 #include "llvm/Transforms/Utils.h"
 
+#ifdef CLING_WITH_ADAPTIVECPP
+#include "hipSYCL/compiler/GlobalsPruningPass.hpp"
+#include "hipSYCL/compiler/SMCPCompatPass.hpp"
+#include "hipSYCL/compiler/sscp/TargetSeparationPass.hpp"
+#endif
+
 //#include "clang/Basic/LangOptions.h"
 //#include "clang/Basic/TargetOptions.h"
 #include "clang/Basic/CharInfo.h"
@@ -511,6 +517,19 @@ void BackendPasses::CreatePasses(int OptLevel, llvm::ModulePassManager& MPM,
   std::optional<PGOOptions> PGOOpt;
   PassBuilder PB(&m_TM, PTO, PGOOpt, &PIC);
 
+#ifdef CLING_WITH_ADAPTIVECPP
+  PB.registerOptimizerLastEPCallback(
+      [](llvm::ModulePassManager& MPM, OptimizationLevel) {
+        MPM.addPass(hipsycl::compiler::SMCPCompatPass{});
+        MPM.addPass(hipsycl::compiler::GlobalsPruningPass{});
+      });
+
+  PB.registerPipelineStartEPCallback(
+      [&](llvm::ModulePassManager& MPM, OptimizationLevel Level) {
+        MPM.addPass(hipsycl::compiler::TargetSeparationPass{""});
+      });
+#endif
+
   // Attempt to load pass plugins and register their callbacks with PB.
   for (auto& PluginFN : m_CGOpts.PassPlugins) {
     auto PassPlugin = PassPlugin::Load(PluginFN);

interpreter/cling/lib/Interpreter/CIFactory.cpp

@@ -1385,6 +1385,19 @@ namespace {
       argvCompile.push_back("-fno-omit-frame-pointer");
     }
 
+#ifdef CLING_WITH_ADAPTIVECPP
+    llvm::SmallString<256> adaptiveCppIncludePath(
+        ADAPTIVE_CPP_HEADER_INSTALL_PATH);
+    if (!llvm::sys::fs::exists(adaptiveCppIncludePath))
+      adaptiveCppIncludePath = ADAPTIVE_CPP_HEADER_BUILD_PATH;
+    argvCompile.push_back("-isystem");
+    argvCompile.push_back(adaptiveCppIncludePath.c_str());
+    argvCompile.push_back("-D__ACPP_ENABLE_LLVM_SSCP_TARGET__");
+    argvCompile.push_back("-Xclang");
+    argvCompile.push_back("-disable-O0-optnone");
+    if (!debuggingEnabled)
+      argvCompile.push_back("-ffp-contract=fast");
+#endif
     // Add host specific includes, -resource-dir if necessary, and -isysroot
     std::string ClingBin = GetExecutablePath(argv[0]);
     AddHostArguments(ClingBin, argvCompile, LLVMDir, COpts);

interpreter/cling/lib/Interpreter/CMakeLists.txt

@@ -23,6 +23,10 @@ set(LIBS
   clangLex
 )
 
+if(adaptivecpp)
+  list(APPEND LIBS acpp-clang acpp-rt)
+endif()
+
 set(LLVM_LINK_COMPONENTS
   analysis
   core
@@ -294,6 +298,9 @@ if (UNIX)
     "
     #define CLING_CXX_INCL \"${CLING_CXX_HEADERS}\"
     #define CLING_INCLUDE_PATHS \"${CLING_INCLUDE_PATHS}\"
+
+    #define ADAPTIVE_CPP_HEADER_BUILD_PATH \"${ADAPTIVE_CPP_HEADER_BUILD_PATH}\"
+    #define ADAPTIVE_CPP_HEADER_INSTALL_PATH \"${ADAPTIVE_CPP_HEADER_INSTALL_PATH}\"
   ")
   if (CMAKE_OSX_SYSROOT)
     # CMAKE_OSX_SYSROOT hardcodes the concrete version of the sdk