aboutsummaryrefslogtreecommitdiffhomepage
diff options
context:
space:
mode:
Diffstat
-rw-r--r--Eigen/src/Core/arch/SYCL/PacketMath.h4
-rw-r--r--cmake/ComputeCppCompilerChecks.cmake50
-rw-r--r--cmake/ComputeCppIRMap.cmake18
-rw-r--r--cmake/FindComputeCpp.cmake33
-rw-r--r--unsupported/test/cxx11_tensor_argmax_sycl.cpp1
5 files changed, 93 insertions, 13 deletions
diff --git a/Eigen/src/Core/arch/SYCL/PacketMath.h b/Eigen/src/Core/arch/SYCL/PacketMath.h
index b11b5af9d..87badc076 100644
--- a/Eigen/src/Core/arch/SYCL/PacketMath.h
+++ b/Eigen/src/Core/arch/SYCL/PacketMath.h
@@ -514,11 +514,11 @@ SYCL_PCMP(eq, cl::sycl::cl_double2)
template <typename T> struct convert_to_integer;
template <> struct convert_to_integer<float> {
- using type = int;
+ using type = std::int32_t;
using packet_type = cl::sycl::cl_int4;
};
template <> struct convert_to_integer<double> {
- using type = long;
+ using type = std::int64_t;
using packet_type = cl::sycl::cl_long2;
};
diff --git a/cmake/ComputeCppCompilerChecks.cmake b/cmake/ComputeCppCompilerChecks.cmake
new file mode 100644
index 000000000..1807485e4
--- /dev/null
+++ b/cmake/ComputeCppCompilerChecks.cmake
@@ -0,0 +1,50 @@
+cmake_minimum_required(VERSION 3.4.3)
+
+if(CMAKE_COMPILER_IS_GNUCXX)
+ if (CMAKE_CXX_COMPILER_VERSION VERSION_LESS 4.8)
+ message(FATAL_ERROR "host compiler - gcc version must be > 4.8")
+ endif()
+elseif ("${CMAKE_CXX_COMPILER_ID}" STREQUAL "Clang")
+ if (${CMAKE_CXX_COMPILER_VERSION} VERSION_LESS 3.6)
+ message(FATAL_ERROR "host compiler - clang version must be > 3.6")
+ endif()
+endif()
+
+if(MSVC)
+ set(ComputeCpp_STL_CHECK_SRC __STL_check)
+ file(WRITE ${CMAKE_CURRENT_BINARY_DIR}/${ComputeCpp_STL_CHECK_SRC}.cpp
+ "#include <ios>\n"
+ "int main() { return 0; }\n")
+ execute_process(
+ COMMAND ${ComputeCpp_DEVICE_COMPILER_EXECUTABLE}
+ ${COMPUTECPP_DEVICE_COMPILER_FLAGS}
+ -isystem ${ComputeCpp_INCLUDE_DIRS}
+ -o ${ComputeCpp_STL_CHECK_SRC}.sycl
+ -c ${ComputeCpp_STL_CHECK_SRC}.cpp
+ WORKING_DIRECTORY ${CMAKE_CURRENT_BINARY_DIR}
+ RESULT_VARIABLE ComputeCpp_STL_CHECK_RESULT
+ ERROR_QUIET
+ OUTPUT_QUIET)
+ if(NOT ${ComputeCpp_STL_CHECK_RESULT} EQUAL 0)
+ # Try disabling compiler version checks
+ execute_process(
+ COMMAND ${ComputeCpp_DEVICE_COMPILER_EXECUTABLE}
+ ${COMPUTECPP_DEVICE_COMPILER_FLAGS}
+ -D_ALLOW_COMPILER_AND_STL_VERSION_MISMATCH
+ -isystem ${ComputeCpp_INCLUDE_DIRS}
+ -o ${ComputeCpp_STL_CHECK_SRC}.cpp.sycl
+ -c ${ComputeCpp_STL_CHECK_SRC}.cpp
+ WORKING_DIRECTORY ${CMAKE_CURRENT_BINARY_DIR}
+ RESULT_VARIABLE ComputeCpp_STL_CHECK_RESULT
+ ERROR_QUIET
+ OUTPUT_QUIET)
+ if(NOT ${ComputeCpp_STL_CHECK_RESULT} EQUAL 0)
+ message(STATUS "Device compiler cannot consume hosted STL headers. Using any parts of the STL will likely result in device compiler errors.")
+ else()
+ message(STATUS "Device compiler does not meet certain STL version requirements. Disabling version checks and hoping for the best.")
+ list(APPEND COMPUTECPP_DEVICE_COMPILER_FLAGS -D_ALLOW_COMPILER_AND_STL_VERSION_MISMATCH)
+ endif()
+ endif()
+ file(REMOVE ${CMAKE_CURRENT_BINARY_DIR}/${ComputeCpp_STL_CHECK_SRC}.cpp
+ ${CMAKE_CURRENT_BINARY_DIR}/${ComputeCpp_STL_CHECK_SRC}.cpp.sycl)
+endif(MSVC)
diff --git a/cmake/ComputeCppIRMap.cmake b/cmake/ComputeCppIRMap.cmake
new file mode 100644
index 000000000..942d91d64
--- /dev/null
+++ b/cmake/ComputeCppIRMap.cmake
@@ -0,0 +1,18 @@
+cmake_minimum_required(VERSION 3.4.3)
+
+# These should match the types of IR output by compute++
+set(IR_MAP_spir bc)
+set(IR_MAP_spir64 bc)
+set(IR_MAP_spir32 bc)
+set(IR_MAP_spirv spv)
+set(IR_MAP_spirv64 spv)
+set(IR_MAP_spirv32 spv)
+set(IR_MAP_aorta-x86_64 o)
+set(IR_MAP_aorta-aarch64 o)
+set(IR_MAP_aorta-rcar-cve o)
+set(IR_MAP_custom-spir64 bc)
+set(IR_MAP_custom-spir32 bc)
+set(IR_MAP_custom-spirv64 spv)
+set(IR_MAP_custom-spirv32 spv)
+set(IR_MAP_ptx64 s)
+set(IR_MAP_amdgcn s)
diff --git a/cmake/FindComputeCpp.cmake b/cmake/FindComputeCpp.cmake
index c926ee292..3cca5150e 100644
--- a/cmake/FindComputeCpp.cmake
+++ b/cmake/FindComputeCpp.cmake
@@ -31,6 +31,7 @@
cmake_minimum_required(VERSION 3.4.3)
include(FindPackageHandleStandardArgs)
+include(ComputeCppIRMap)
set(COMPUTECPP_USER_FLAGS "" CACHE STRING "User flags for compute++")
separate_arguments(COMPUTECPP_USER_FLAGS)
@@ -64,11 +65,13 @@ endif()
find_program(ComputeCpp_DEVICE_COMPILER_EXECUTABLE compute++
HINTS ${computecpp_host_find_hint}
- PATH_SUFFIXES bin)
+ PATH_SUFFIXES bin
+ NO_SYSTEM_ENVIRONMENT_PATH)
find_program(ComputeCpp_INFO_EXECUTABLE computecpp_info
HINTS ${computecpp_host_find_hint}
- PATH_SUFFIXES bin)
+ PATH_SUFFIXES bin
+ NO_SYSTEM_ENVIRONMENT_PATH)
find_library(COMPUTECPP_RUNTIME_LIBRARY
NAMES ComputeCpp ComputeCpp_vs2015
@@ -77,7 +80,7 @@ find_library(COMPUTECPP_RUNTIME_LIBRARY
DOC "ComputeCpp Runtime Library")
find_library(COMPUTECPP_RUNTIME_LIBRARY_DEBUG
- NAMES ComputeCpp ComputeCpp_vs2015_d
+ NAMES ComputeCpp_d ComputeCpp ComputeCpp_vs2015_d
HINTS ${computecpp_find_hint}
PATH_SUFFIXES lib
DOC "ComputeCpp Debug Runtime Library")
@@ -112,7 +115,7 @@ else()
if (COMPUTECPP_PLATFORM_IS_SUPPORTED)
message(STATUS "platform - your system can support ComputeCpp")
else()
- message(WARNING "platform - your system CANNOT support ComputeCpp")
+ message(STATUS "platform - your system is not officially supported")
endif()
endif()
endif()
@@ -151,6 +154,8 @@ endif()
list(APPEND COMPUTECPP_DEVICE_COMPILER_FLAGS -sycl-target ${COMPUTECPP_BITCODE})
message(STATUS "compute++ flags - ${COMPUTECPP_DEVICE_COMPILER_FLAGS}")
+include(ComputeCppCompilerChecks)
+
if(NOT TARGET OpenCL::OpenCL)
add_library(OpenCL::OpenCL UNKNOWN IMPORTED)
set_target_properties(OpenCL::OpenCL PROPERTIES
@@ -163,7 +168,7 @@ if(NOT TARGET ComputeCpp::ComputeCpp)
add_library(ComputeCpp::ComputeCpp UNKNOWN IMPORTED)
set_target_properties(ComputeCpp::ComputeCpp PROPERTIES
IMPORTED_LOCATION_DEBUG "${COMPUTECPP_RUNTIME_LIBRARY_DEBUG}"
- IMPORTED_LOCATION_RELWITHDEBINFO "${COMPUTECPP_RUNTIME_LIBRARY_DEBUG}"
+ IMPORTED_LOCATION_RELWITHDEBINFO "${COMPUTECPP_RUNTIME_LIBRARY}"
IMPORTED_LOCATION "${COMPUTECPP_RUNTIME_LIBRARY}"
INTERFACE_INCLUDE_DIRECTORIES "${ComputeCpp_INCLUDE_DIRS}"
INTERFACE_LINK_LIBRARIES "OpenCL::OpenCL"
@@ -200,7 +205,7 @@ define_property(
####################
#
# Adds a custom target for running compute++ and adding a dependency for the
-# resulting integration header.
+# resulting integration header and kernel binary.
#
# TARGET : Name of the target.
# SOURCE : Source file to be compiled.
@@ -230,6 +235,7 @@ function(__build_ir)
# using the same source file will be generated with a different rule.
set(baseSyclName ${CMAKE_CURRENT_BINARY_DIR}/${SDK_BUILD_IR_TARGET}_${sourceFileName})
set(outputSyclFile ${baseSyclName}.sycl)
+ set(outputDeviceFile ${baseSyclName}.${IR_MAP_${COMPUTECPP_BITCODE}})
set(depFileName ${baseSyclName}.sycl.d)
set(include_directories "$<TARGET_PROPERTY:${SDK_BUILD_IR_TARGET},INCLUDE_DIRECTORIES>")
@@ -273,7 +279,9 @@ function(__build_ir)
get_target_property(target_libraries ${SDK_BUILD_IR_TARGET} LINK_LIBRARIES)
if(target_libraries)
foreach(library ${target_libraries})
- list(APPEND ir_dependencies ${library})
+ if(TARGET ${library})
+ list(APPEND ir_dependencies ${library})
+ endif()
endforeach()
endif()
@@ -281,19 +289,20 @@ function(__build_ir)
# CMake throws an error if it is unsupported by the generator (i. e. not ninja)
if((NOT CMAKE_VERSION VERSION_LESS 3.7.0) AND
CMAKE_GENERATOR MATCHES "Ninja")
- file(RELATIVE_PATH relOutputFile ${CMAKE_BINARY_DIR} ${outputSyclFile})
+ file(RELATIVE_PATH relOutputFile ${CMAKE_BINARY_DIR} ${outputDeviceFile})
set(generate_depfile -MMD -MF ${depFileName} -MT ${relOutputFile})
set(enable_depfile DEPFILE ${depFileName})
endif()
# Add custom command for running compute++
add_custom_command(
- OUTPUT ${outputSyclFile}
+ OUTPUT ${outputDeviceFile} ${outputSyclFile}
COMMAND ${ComputeCpp_DEVICE_COMPILER_EXECUTABLE}
${COMPUTECPP_DEVICE_COMPILER_FLAGS}
${generated_include_directories}
${generated_compile_definitions}
- -o ${outputSyclFile}
+ -sycl-ih ${outputSyclFile}
+ -o ${outputDeviceFile}
-c ${SDK_BUILD_IR_SOURCE}
${generate_depfile}
DEPENDS ${ir_dependencies}
@@ -308,7 +317,7 @@ function(__build_ir)
if(NOT MSVC)
# Add a custom target for the generated integration header
- add_custom_target(${headerTargetName} DEPENDS ${outputSyclFile})
+ add_custom_target(${headerTargetName} DEPENDS ${outputDeviceFile} ${outputSyclFile})
add_dependencies(${SDK_BUILD_IR_TARGET} ${headerTargetName})
endif()
@@ -392,6 +401,8 @@ function(add_sycl_to_target)
${ARGN}
)
+ set_target_properties(${SDK_ADD_SYCL_TARGET} PROPERTIES LINKER_LANGUAGE CXX)
+
# If the CXX compiler is set to compute++ enable the driver.
get_filename_component(cmakeCxxCompilerFileName "${CMAKE_CXX_COMPILER}" NAME)
if("${cmakeCxxCompilerFileName}" STREQUAL "compute++")
diff --git a/unsupported/test/cxx11_tensor_argmax_sycl.cpp b/unsupported/test/cxx11_tensor_argmax_sycl.cpp
index 41ea3cf7b..7ac71286e 100644
--- a/unsupported/test/cxx11_tensor_argmax_sycl.cpp
+++ b/unsupported/test/cxx11_tensor_argmax_sycl.cpp
@@ -16,6 +16,7 @@
#define EIGEN_DEFAULT_DENSE_INDEX_TYPE int64_t
#define EIGEN_USE_SYCL
+#define EIGEN_HAS_CONSTEXPR 1
#include "main.h"
generated by cgit v1.2.3 (git 2.39.1) at 2024-11-27 13:57:09 +0000