[SYCL] Minor cleanup to improve build times and also fix failed tests

.github/workflows/sycl-cpu-ci.yml

@@ -0,0 +1,82 @@
+# .github/workflows/sycl-ci.yml
+#
+# Builds and tests ExchCXX SYCL backend on CPU using Intel oneAPI.
+#
+# Triggers:
+#   1. PR title contains "[SYCL]"
+#   2. "SYCL" label is added to a PR
+#   3. Manually via "Run workflow" button
+
+name: SYCL CI
+
+on:
+  pull_request:
+    types: [opened, edited, synchronize, labeled]
+
+  workflow_dispatch:
+    inputs:
+      branch:
+        description: 'Branch or commit SHA to test'
+        required: false
+        default: ''
+
+jobs:
+  build-and-test-sycl:
+    if: >
+      github.event_name == 'workflow_dispatch' ||
+      (github.event_name == 'pull_request' && github.event.action == 'labeled' && github.event.label.name == 'SYCL') ||
+      (github.event_name == 'pull_request' && github.event.action != 'labeled' && contains(github.event.pull_request.title, '[SYCL]'))
+    runs-on: ubuntu-latest
+    timeout-minutes: 45
+
+    steps:
+      - name: Checkout
+        uses: actions/checkout@v4
+        with:
+          ref: ${{ github.event.inputs.branch || github.event.pull_request.head.sha || github.sha }}
+
+      - name: Add Intel oneAPI apt repository
+        run: |
+          wget -qO- https://apt.repos.intel.com/intel-gpg-keys/GPG-PUB-KEY-INTEL-SW-PRODUCTS.PUB \
+            | sudo gpg --dearmor -o /usr/share/keyrings/oneapi-archive-keyring.gpg
+          echo "deb [signed-by=/usr/share/keyrings/oneapi-archive-keyring.gpg] https://apt.repos.intel.com/oneapi all main" \
+            | sudo tee /etc/apt/sources.list.d/oneAPI.list
+          sudo apt-get update
+
+      - name: Install Intel oneAPI DPC++/C++ compiler and OpenCL CPU runtime
+        run: |
+          sudo apt-get install -y \
+            intel-oneapi-compiler-dpcpp-cpp \
+            intel-oneapi-runtime-opencl \
+            ninja-build
+
+      - name: Configure
+        run: |
+          source /opt/intel/oneapi/setvars.sh
+
+          cmake -H. -Bbuild_sycl_cpu -G Ninja \
+            -DCMAKE_C_COMPILER=icx \
+            -DCMAKE_CXX_COMPILER=icpx \
+            -DEXCHCXX_ENABLE_SYCL=ON \
+            -DEXCHCXX_SYCL_TARGET=spir64_x86_64 \
+            -DBUILD_SHARED_LIBS=ON \
+            -DEXCHCXX_ENABLE_TESTS=ON \
+            -DCMAKE_POSITION_INDEPENDENT_CODE=ON \
+            -DCMAKE_BUILD_TYPE=Release
+
+      - name: Build
+        run: |
+          source /opt/intel/oneapi/setvars.sh
+          cmake --build build_sycl_cpu -j $(nproc)
+
+      - name: List SYCL devices
+        run: |
+          source /opt/intel/oneapi/setvars.sh
+          sycl-ls || true
+
+      - name: Test
+        run: |
+          source /opt/intel/oneapi/setvars.sh
+          export ONEAPI_DEVICE_SELECTOR=opencl:cpu
+          cd build_sycl_cpu
+          ctest --output-on-failure -j $(nproc)

CMakeLists.txt

@@ -30,8 +30,20 @@ endif()
 
 
 if(EXCHCXX_ENABLE_SYCL)
-  # e.g. intel_gpu_pvc | nvidia_gpu_sm_80 | nvidia_gpu_sm_90 | amd_gpu_gfx90a | amd_gpu_gfx942
+  # e.g. intel_gpu_pvc | spir64_x86_64 | nvidia_gpu_sm_80 | nvidia_gpu_sm_90 | amd_gpu_gfx90a | amd_gpu_gfx942
   set(EXCHCXX_SYCL_TARGET "" CACHE STRING "Alias for -fsycl-targets (see Users Manual)")
+
+  include( CheckCCompilerFlag )
+  include( CheckCXXCompilerFlag )
+  check_c_compiler_flag  ( "-fp-model=precise" EXCHCXX_C_HAS_FP_MODEL_PRECISE   )
+  check_cxx_compiler_flag( "-fp-model=precise" EXCHCXX_CXX_HAS_FP_MODEL_PRECISE )
+
+  if( EXCHCXX_C_HAS_FP_MODEL_PRECISE )
+    set( CMAKE_C_FLAGS "${CMAKE_C_FLAGS} -fp-model=precise" )
+  endif()
+  if( EXCHCXX_CXX_HAS_FP_MODEL_PRECISE )
+    set( CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -fp-model=precise" )
+  endif()
 endif()
 
 
@@ -89,6 +101,14 @@ else()
   set( BUILD_TESTING OFF CACHE BOOL "" FORCE )
 
   FetchContent_MakeAvailable( libxc )
+
+  #Note: Missing this would cause test failures
+  if(CMAKE_C_COMPILER_ID STREQUAL "IntelLLVM")
+    target_compile_options(xc PRIVATE
+      $<$<COMPILE_LANGUAGE:C>:-fp-model=precise>
+    )
+  endif()
+
   add_library( Libxc::xc ALIAS xc )
   target_include_directories( xc
     PUBLIC

include/exchcxx/exchcxx_config.hpp.in

@@ -62,4 +62,5 @@
 
 #ifdef EXCHCXX_ENABLE_SYCL
   #include <sycl/sycl.hpp>
+  namespace syclex = sycl::ext::oneapi;
 #endif

include/exchcxx/impl/builtin/kernels/deorbitalized.hpp

@@ -142,7 +142,7 @@ struct kernel_traits<Deorbitalized<XCEF,KEDF>> {
                          double& v2lapl2, double& v2lapltau, double& v2tau2 ) {
     #if defined(__CUDACC__) || defined(__HIPCC__)
     printf("eval_vxc_fxc_unpolar not implemented for deorbitalized kernels\n");
-    #elif defined(__SYCL_DEVICE_ONLY__) || defined(EXCHCXX_ENABLE_SYCL)
+    #elif defined(__SYCL_DEVICE_ONLY__) && defined(EXCHCXX_ENABLE_SYCL)
     sycl::ext::oneapi::experimental::printf("eval_vxc_fxc_unpolar not implemented for deorbitalized kernels\n");
     #else
     unused(rho, sigma, lapl, tau, vrho, vsigma, vlapl, vtau, v2rho2, v2rhosigma, v2rholapl, v2rhotau, v2sigma2, v2sigmalapl, v2sigmatau, v2lapl2, v2lapltau, v2tau2);
@@ -173,7 +173,7 @@ struct kernel_traits<Deorbitalized<XCEF,KEDF>> {
                        double& v2tau2_aa, double& v2tau2_ab, double& v2tau2_bb ) {
     #if defined(__CUDACC__) || defined(__HIPCC__)
     printf("eval_vxc_fxc_polar not implemented for deorbitalized kernels\n");
-    #elif defined(__SYCL_DEVICE_ONLY__) || defined(EXCHCXX_ENABLE_SYCL)
+    #elif defined(__SYCL_DEVICE_ONLY__) && defined(EXCHCXX_ENABLE_SYCL)
     sycl::ext::oneapi::experimental::printf("eval_vxc_fxc_polar not implemented for deorbitalized kernels\n");
     #else
       unused(rho_a, rho_b, sigma_aa, sigma_ab, sigma_bb, lapl_a, lapl_b, tau_a, tau_b, vrho_a, vrho_b, vsigma_aa, vsigma_ab, vsigma_bb, vlapl_a, vlapl_b, vtau_a, vtau_b, v2rho2_aa, v2rho2_ab, v2rho2_bb, v2rhosigma_a_aa, v2rhosigma_a_ab, v2rhosigma_a_bb, v2rhosigma_b_aa, v2rhosigma_b_ab, v2rhosigma_b_bb, v2rholapl_a_a, v2rholapl_a_b, v2rholapl_b_a, v2rholapl_b_b, v2rhotau_a_a, v2rhotau_a_b, v2rhotau_b_a, v2rhotau_b_b, v2sigma2_aa_aa, v2sigma2_aa_ab, v2sigma2_aa_bb, v2sigma2_ab_ab, v2sigma2_ab_bb, v2sigma2_bb_bb, v2sigmalapl_aa_a, v2sigmalapl_aa_b, v2sigmalapl_ab_a, v2sigmalapl_ab_b, v2sigmalapl_bb_a, v2sigmalapl_bb_b, v2sigmatau_aa_a, v2sigmatau_aa_b, v2sigmatau_ab_a, v2sigmatau_ab_b, v2sigmatau_bb_a, v2sigmatau_bb_b, v2lapl2_aa, v2lapl2_ab, v2lapl2_bb, v2lapltau_a_a, v2lapltau_a_b, v2lapltau_b_a, v2lapltau_b_b, v2tau2_aa, v2tau2_ab, v2tau2_bb);
@@ -189,7 +189,7 @@ struct kernel_traits<Deorbitalized<XCEF,KEDF>> {
                      double& v2lapl2, double& v2lapltau, double& v2tau2 ) {
     #if defined(__CUDACC__) || defined(__HIPCC__)
     printf("eval_fxc_unpolar not implemented for deorbitalized kernels\n");
-    #elif defined(__SYCL_DEVICE_ONLY__) || defined(EXCHCXX_ENABLE_SYCL)
+    #elif defined(__SYCL_DEVICE_ONLY__) && defined(EXCHCXX_ENABLE_SYCL)
     sycl::ext::oneapi::experimental::printf("eval_fxc_unpolar not implemented for deorbitalized kernels\n");
     #else
     unused(rho, sigma, lapl, tau, v2rho2, v2rhosigma, v2rholapl, v2rhotau, v2sigma2, v2sigmalapl, v2sigmatau, v2lapl2, v2lapltau, v2tau2);
@@ -217,7 +217,7 @@ struct kernel_traits<Deorbitalized<XCEF,KEDF>> {
                    double& v2tau2_aa, double& v2tau2_ab, double& v2tau2_bb ) {
     #if defined(__CUDACC__) || defined(__HIPCC__)
     printf("eval_fxc_polar not implemented for deorbitalized kernels\n");
-    #elif defined(__SYCL_DEVICE_ONLY__) || defined(EXCHCXX_ENABLE_SYCL)
+    #elif defined(__SYCL_DEVICE_ONLY__) && defined(EXCHCXX_ENABLE_SYCL)
     sycl::ext::oneapi::experimental::printf("eval_fxc_polar not implemented for deorbitalized kernels\n");
     #else
     unused(rho_a, rho_b, sigma_aa, sigma_ab, sigma_bb, lapl_a, lapl_b, tau_a, tau_b, v2rho2_aa, v2rho2_ab, v2rho2_bb, v2rhosigma_a_aa, v2rhosigma_a_ab, v2rhosigma_a_bb, v2rhosigma_b_aa, v2rhosigma_b_ab, v2rhosigma_b_bb, v2rholapl_a_a, v2rholapl_a_b, v2rholapl_b_a, v2rholapl_b_b, v2rhotau_a_a, v2rhotau_a_b, v2rhotau_b_a, v2rhotau_b_b, v2sigma2_aa_aa, v2sigma2_aa_ab, v2sigma2_aa_bb, v2sigma2_ab_ab, v2sigma2_ab_bb, v2sigma2_bb_bb, v2sigmalapl_aa_a, v2sigmalapl_aa_b, v2sigmalapl_ab_a, v2sigmalapl_ab_b, v2sigmalapl_bb_a, v2sigmalapl_bb_b, v2sigmatau_aa_a, v2sigmatau_aa_b, v2sigmatau_ab_a, v2sigmatau_ab_b, v2sigmatau_bb_a, v2sigmatau_bb_b, v2lapl2_aa, v2lapl2_ab, v2lapl2_bb, v2lapltau_a_a, v2lapltau_a_b, v2lapltau_b_a, v2lapltau_b_b, v2tau2_aa, v2tau2_ab, v2tau2_bb);

include/exchcxx/impl/builtin/util.hpp

@@ -111,9 +111,16 @@ namespace sm = std;
 template <typename T>
 SAFE_INLINE(auto) sqrt( T x ) { return sm::sqrt(x); }
 
-// Apparently C / C++ cbrt differ ever-so-slightly
+// Apparently C / C++ cbrt differ ever-so-slightly. ::cbrt is used instead of
+// sm::cbrt on Host/CUDA/HIP for that reason; SYCL is the exception because
+// ::cbrt (host libm) has no SPIR-V device image, so sycl::cbrt is required
+// there for AoT device compilation to link.
 template <typename T>
+#ifdef EXCHCXX_ENABLE_SYCL
+SAFE_INLINE(double) cbrt( T x ) { return sm::cbrt(x); }
+#else
 SAFE_INLINE(double) cbrt( T x ) { return ::cbrt(x); }
+#endif
 template <typename T>
 SAFE_INLINE(auto) log( T x ) { return sm::log(x); }
 template <typename T>