From 052426b824038bbee1c1c48c3df65dfccd79ae24 Mon Sep 17 00:00:00 2001 From: a-doumoulakis Date: Fri, 5 May 2017 19:26:27 +0100 Subject: Add support for triSYCL Eigen is now able to use triSYCL with EIGEN_SYCL_TRISYCL and TRISYCL_INCLUDE_DIR options Fix contraction kernel with correct nd_item dimension --- unsupported/Eigen/CXX11/Tensor | 2 +- .../Eigen/CXX11/src/Tensor/TensorContractionSycl.h | 2 +- unsupported/test/CMakeLists.txt | 59 ++++++++++++---------- 3 files changed, 35 insertions(+), 28 deletions(-) (limited to 'unsupported') diff --git a/unsupported/Eigen/CXX11/Tensor b/unsupported/Eigen/CXX11/Tensor index 39916092b..5d71a9c25 100644 --- a/unsupported/Eigen/CXX11/Tensor +++ b/unsupported/Eigen/CXX11/Tensor @@ -19,7 +19,7 @@ #undef isnan #undef isinf #undef isfinite -#include +#include #include #include #include diff --git a/unsupported/Eigen/CXX11/src/Tensor/TensorContractionSycl.h b/unsupported/Eigen/CXX11/src/Tensor/TensorContractionSycl.h index 5b4c3c5bd..e6840bc87 100644 --- a/unsupported/Eigen/CXX11/src/Tensor/TensorContractionSycl.h +++ b/unsupported/Eigen/CXX11/src/Tensor/TensorContractionSycl.h @@ -195,7 +195,7 @@ typename HostExpr::Index LocalThreadSizeM, typename HostExpr::Index LocalThreadS m_j_strides(m_j_strides_), m_right_nocontract_strides(m_right_nocontract_strides_), left_tuple_of_accessors(left_tuple_of_accessors_), right_tuple_of_accessors(right_tuple_of_accessors_), dev(dev_){} - void operator()(cl::sycl::nd_item<1> itemID) { + void operator()(cl::sycl::nd_item<2> itemID) { typedef typename Eigen::TensorSycl::internal::ConvertToDeviceExpression::Type DevExpr; typedef typename Eigen::TensorSycl::internal::ConvertToDeviceExpression::Type LHSDevExpr; typedef typename Eigen::TensorSycl::internal::ConvertToDeviceExpression::Type RHSDevExpr; diff --git a/unsupported/test/CMakeLists.txt b/unsupported/test/CMakeLists.txt index cdf151f15..9a9124640 100644 --- a/unsupported/test/CMakeLists.txt +++ b/unsupported/test/CMakeLists.txt @@ -152,33 +152,40 @@ endif() if(EIGEN_TEST_CXX11) if(EIGEN_TEST_SYCL) - ei_add_test_sycl(cxx11_tensor_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_forced_eval_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_broadcast_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_device_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_reduction_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_morphing_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_shuffling_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_padding_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_builtins_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_contract_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_concatenation_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_reverse_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_convolution_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_striding_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_chipping_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_layout_swap_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_inflation_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_generator_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_patch_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_image_patch_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_volume_patcP_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_argmax_sycl "-std=c++11") - ei_add_test_sycl(cxx11_tensor_custom_op_sycl "-std=c++11") + if(EIGEN_SYCL_TRISYCL) + set(CMAKE_CXX_STANDARD 14) + set(STD_CXX_FLAG "-std=c++1z") + else(EIGEN_SYCL_TRISYCL) + # It should be safe to always run these tests as there is some fallback code for + # older compiler that don't support cxx11. + set(CMAKE_CXX_STANDARD 11) + set(STD_CXX_FLAG "-std=c++11") + endif(EIGEN_SYCL_TRISYCL) + + ei_add_test_sycl(cxx11_tensor_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_forced_eval_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_broadcast_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_device_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_reduction_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_morphing_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_shuffling_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_padding_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_builtins_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_contract_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_concatenation_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_reverse_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_convolution_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_striding_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_chipping_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_layout_swap_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_inflation_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_generator_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_patch_sycl ${STD_CXX_FLAG}) + #ei_add_test_sycl(cxx11_tensor_image_patch_sycl ${STD_CXX_FLAG}) + #ei_add_test_sycl(cxx11_tensor_volume_patch_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_argmax_sycl ${STD_CXX_FLAG}) + ei_add_test_sycl(cxx11_tensor_custom_op_sycl ${STD_CXX_FLAG}) endif(EIGEN_TEST_SYCL) - # It should be safe to always run these tests as there is some fallback code for - # older compiler that don't support cxx11. - set(CMAKE_CXX_STANDARD 11) ei_add_test(cxx11_eventcount "-pthread" "${CMAKE_THREAD_LIBS_INIT}") ei_add_test(cxx11_runqueue "-pthread" "${CMAKE_THREAD_LIBS_INIT}") -- cgit v1.2.3