Commit: fee6dd197242e5bd9b46f7873ce5fd08281e857f Author: Xavier Hallade Date: Tue Apr 19 14:44:10 2022 +0200 Branches: cycles_oneapi https://developer.blender.org/rBfee6dd197242e5bd9b46f7873ce5fd08281e857f
Cycles: add intel/llvm sycl compiler support to Intel GPUs backend on Windows. tested with release 2021-12: https://github.com/intel/llvm/releases/tag/2021-12 with the below workaround for intel/llvm headers: sed -i 's/(detail::declptr/(::sycl::detail::declptr/g' ./build/install/include/sycl/CL/sycl/nd_item.hpp =================================================================== M build_files/cmake/Modules/FindSYCL.cmake M intern/cycles/kernel/CMakeLists.txt M intern/cycles/kernel/device/oneapi/compat.h M intern/cycles/kernel/device/oneapi/kernel.cpp =================================================================== diff --git a/build_files/cmake/Modules/FindSYCL.cmake b/build_files/cmake/Modules/FindSYCL.cmake index 0b42da8cf4f..6a35ff360c4 100644 --- a/build_files/cmake/Modules/FindSYCL.cmake +++ b/build_files/cmake/Modules/FindSYCL.cmake @@ -4,7 +4,7 @@ # - Find SYCL library # Find the native SYCL header and libraries needed by oneAPI implementation # This module defines -# SYCL_DPCPP_COMPILER, compiler from oneAPI toolkit, which will be used for compilation of SYCL code +# SYCL_COMPILER, compiler which will be used for compilation of SYCL code # SYCL_LIBRARY, libraries to link against in order to use SYCL. # SYCL_INCLUDE_DIR, directories where SYCL headers can be found # SYCL_ROOT_DIR, The base directory to search for SYCL files. @@ -23,7 +23,7 @@ SET(_sycl_search_dirs C:/Program\ Files\ \(x86\)/Intel/oneAPI/compiler/latest/windows ) -FIND_PROGRAM(SYCL_DPCPP_COMPILER +FIND_PROGRAM(SYCL_COMPILER NAMES dpcpp HINTS @@ -32,7 +32,7 @@ FIND_PROGRAM(SYCL_DPCPP_COMPILER bin ) -FIND_LIBRARY(_SYCL_LIBRARY +FIND_LIBRARY(SYCL_LIBRARY NAMES sycl HINTS @@ -41,7 +41,7 @@ FIND_LIBRARY(_SYCL_LIBRARY lib64 lib ) -FIND_PATH(_SYCL_INCLUDE_DIR +FIND_PATH(SYCL_INCLUDE_DIR NAMES CL/sycl.hpp HINTS @@ -53,23 +53,15 @@ FIND_PATH(_SYCL_INCLUDE_DIR INCLUDE(FindPackageHandleStandardArgs) -FIND_PACKAGE_HANDLE_STANDARD_ARGS(SYCL DEFAULT_MSG _SYCL_LIBRARY _SYCL_INCLUDE_DIR) +FIND_PACKAGE_HANDLE_STANDARD_ARGS(SYCL DEFAULT_MSG SYCL_LIBRARY SYCL_INCLUDE_DIR) IF(SYCL_FOUND) - SET(SYCL_LIBRARY ${_SYCL_LIBRARY}) - - get_filename_component(_SYCL_INCLUDE_PARENT_DIR ${_SYCL_INCLUDE_DIR} DIRECTORY) - - SET(SYCL_INCLUDE_DIR ${_SYCL_INCLUDE_DIR} ${_SYCL_INCLUDE_PARENT_DIR}) + get_filename_component(_SYCL_INCLUDE_PARENT_DIR ${SYCL_INCLUDE_DIR} DIRECTORY) + SET(SYCL_INCLUDE_DIR ${SYCL_INCLUDE_DIR} ${_SYCL_INCLUDE_PARENT_DIR}) ELSE() SET(SYCL_SYCL_FOUND FALSE) ENDIF() MARK_AS_ADVANCED( - SYCL_LIBRARY - SYCL_INCLUDE_DIR - SYCL_DPCPP_COMPILER - _SYCL_INCLUDE_DIR _SYCL_INCLUDE_PARENT_DIR - _SYCL_LIBRARY ) diff --git a/intern/cycles/kernel/CMakeLists.txt b/intern/cycles/kernel/CMakeLists.txt index f2e68287f4c..1464f66e512 100644 --- a/intern/cycles/kernel/CMakeLists.txt +++ b/intern/cycles/kernel/CMakeLists.txt @@ -719,7 +719,7 @@ if(WITH_CYCLES_DEVICE_ONEAPI) ${SRC_UTIL_HEADERS} ) - set(dpcpp_flags + set(sycl_compiler_flags ${CMAKE_CURRENT_SOURCE_DIR}/${SRC_KERNEL_DEVICE_ONEAPI} -fsycl -fsycl-unnamed-lambda @@ -736,25 +736,25 @@ if(WITH_CYCLES_DEVICE_ONEAPI) ${ONEAPI_DPCPP_FLAGS} ) if (WITH_CYCLES_ONEAPI_SYCL_HOST_ENABLED) - list(APPEND dpcpp_flags -DWITH_ONEAPI_SYCL_HOST_ENABLED) + list(APPEND sycl_compiler_flags -DWITH_ONEAPI_SYCL_HOST_ENABLED) endif() if (WITH_CYCLES_ONEAPI_BINARIES) set (CYCLES_ONEAPI_GPU_COMPILATION_OPTIONS "-internal_options '-ze-opt-large-register-file -ze-opt-regular-grf-kernel integrator_intersect'") - list(APPEND dpcpp_flags + list(APPEND sycl_compiler_flags -fsycl-targets=spir64_gen -Xsycl-target-backend=spir64_gen "-device ${CYCLES_ONEAPI_AOT_TARGETS} ${CYCLES_ONEAPI_GPU_COMPILATION_OPTIONS}") endif() if(WITH_NANOVDB) - list(APPEND dpcpp_flags + list(APPEND sycl_compiler_flags -DWITH_NANOVDB -I"${NANOVDB_INCLUDE_DIR}") endif() - get_filename_component(dpcpp_root ${SYCL_DPCPP_COMPILER} DIRECTORY) + get_filename_component(sycl_compiler_root ${SYCL_COMPILER} DIRECTORY) + get_filename_component(sycl_compiler_compiler_name ${SYCL_COMPILER} NAME_WE) if(WIN32) - list(APPEND dpcpp_flags - /EHsc + list(APPEND sycl_compiler_flags -fms-extensions -fms-compatibility -D_WINDLL @@ -763,16 +763,31 @@ if(WITH_CYCLES_DEVICE_ONEAPI) -D_WINDOWS -DONEAPI_EXPORT) - add_custom_command( - OUTPUT ${cycles_kernel_oneapi_lib} - COMMAND "${dpcpp_root}/../../env/vars.bat" - COMMAND ${SYCL_DPCPP_COMPILER} ${dpcpp_flags} - DEPENDS ${cycles_oneapi_kernel_sources}) + if(sycl_compiler_compiler_name MATCHES "dpcpp") + add_custom_command( + OUTPUT ${cycles_kernel_oneapi_lib} + COMMAND "${sycl_compiler_root}/../../env/vars.bat" + COMMAND ${SYCL_COMPILER} ${sycl_compiler_flags} + DEPENDS ${cycles_oneapi_kernel_sources}) + else() + string(REPLACE /Redist/ /Tools/ MSVC_TOOLS_DIR ${MSVC_REDIST_DIR}) + list(APPEND sycl_compiler_flags + -L "${MSVC_TOOLS_DIR}/lib/x64" + -L "${WINDOWS_KITS_DIR}/Lib/${CMAKE_VS_WINDOWS_TARGET_PLATFORM_VERSION}/um/x64" + -L "${WINDOWS_KITS_DIR}/Lib/${CMAKE_VS_WINDOWS_TARGET_PLATFORM_VERSION}/ucrt/x64") + add_custom_command( + OUTPUT ${cycles_kernel_oneapi_lib} + COMMAND ${CMAKE_COMMAND} -E env + "LIB=${sycl_compiler_root}/../lib;${LIB}" + "PATH=${sycl_compiler_root}/../lib/ocloc;${sycl_compiler_root};${PATH}" + ${SYCL_COMPILER} ${sycl_compiler_flags} + DEPENDS ${cycles_oneapi_kernel_sources}) + endif() else() - list(APPEND dpcpp_flags -fPIC) + list(APPEND sycl_compiler_flags -fPIC) add_custom_command( OUTPUT ${cycles_kernel_oneapi_lib} - COMMAND bash -c \"source ${dpcpp_root}/../../env/vars.sh&&${SYCL_DPCPP_COMPILER} ${dpcpp_flags}\" + COMMAND bash -c \"source ${sycl_compiler_root}/../../env/vars.sh&&${SYCL_COMPILER} ${sycl_compiler_flags}\" DEPENDS ${cycles_oneapi_kernel_sources}) endif() diff --git a/intern/cycles/kernel/device/oneapi/compat.h b/intern/cycles/kernel/device/oneapi/compat.h index 0c0ec827a71..d737f5a8029 100644 --- a/intern/cycles/kernel/device/oneapi/compat.h +++ b/intern/cycles/kernel/device/oneapi/compat.h @@ -192,6 +192,7 @@ ccl_always_inline float3 make_float3(float x) #include "util/half.h" #include "util/types.h" + // NOTE(sirgienko) Declaring these functions after types headers is very important because they // include oneAPI headers, which transitively include math.h headers which will cause redefintions // of the math defines because math.h also uses them and having them defined before math.h include diff --git a/intern/cycles/kernel/device/oneapi/kernel.cpp b/intern/cycles/kernel/device/oneapi/kernel.cpp index a80e04edc87..8d480566dea 100644 --- a/intern/cycles/kernel/device/oneapi/kernel.cpp +++ b/intern/cycles/kernel/device/oneapi/kernel.cpp @@ -10,6 +10,7 @@ # include <set> # include <level_zero/ze_api.h> +# include <CL/sycl.hpp> # include <ext/oneapi/backend/level_zero.hpp> # include "kernel/device/oneapi/compat.h" @@ -185,7 +186,7 @@ bool oneapi_trigger_runtime_compilation(SyclQueue *queue_) try { queue->submit([&](sycl::handler &cgh) { - sycl::accessor A_acc(A, cgh, sycl::read_only, sycl::no_init); + sycl::accessor A_acc(A, cgh, sycl::read_only); sycl::accessor B_acc(B, cgh, sycl::write_only, sycl::no_init); cgh.parallel_for(N, [=](sycl::id<1> idx) { B_acc[idx] = A_acc[idx] + idx.get(0); }); _______________________________________________ Bf-blender-cvs mailing list [email protected] List details, subscription details or unsubscribe: https://lists.blender.org/mailman/listinfo/bf-blender-cvs
