diff --git a/README.md b/README.md index b7755bf1b..ed61aa946 100644 --- a/README.md +++ b/README.md @@ -425,12 +425,12 @@ by AdaptiveCpp. cd build export CC=[path/to/system/clang] export CXX=[path/to/AdaptiveCpp/install/bin/syclcc] -export HIPSYCL_TARGETS=[compilation_flow:target] # (e.g. cuda:sm_75) +export ACPP_TARGETS=[compilation_flow:target] # (e.g. cuda:sm_75) cmake -GNinja ../ -DAdaptiveCpp_DIR=/path/to/AdaptiveCpp/install/lib/cmake/AdaptiveCpp \ - -DSYCL_COMPILER=adaptivecpp -DHIPSYCL_TARGETS=$HIPSYCL_TARGETS + -DSYCL_COMPILER=adaptivecpp -DHIPSYCL_TARGETS=$ACPP_TARGETS ninja ``` -To build for other than the default devices (`omp`), set the `HIPSYCL_TARGETS` environment +To build for other than the default devices (`omp`), set the `ACPP_TARGETS` environment variable or specify `-DHIPSYCL_TARGETS` as [documented](https://github.com/AdaptiveCpp/AdaptiveCpp/blob/develop/doc/using-hipsycl.md). The available backends are the ones built with AdaptiveCpp in the first place. @@ -477,7 +477,7 @@ Some of the supported options are: |---|---|---| | `BLAS_ENABLE_TESTING` | `ON`/`OFF` | Set it to `OFF` to avoid building the tests (`ON` is the default value) | | `BLAS_ENABLE_BENCHMARK` | `ON`/`OFF` | Set it to `OFF` to avoid building the benchmarks (`ON` is the default value) | -| `SYCL_COMPILER` | name | Used to determine which SYCL implementation to use. By default, the first implementation found is used. Supported values are: `dpcpp`, `hipsycl` and `computecpp`*(deprecated)*. | +| `SYCL_COMPILER` | name | Used to determine which SYCL implementation to use. By default, the first implementation found is used. Supported values are: `dpcpp`, `adaptivecpp` and `computecpp`*(deprecated)*. | | `TUNING_TARGET` | name | By default, this flag is set to `DEFAULT_CPU` to restrict any device specific compiler optimizations. Use this flag to tune the code for a target (**highly recommended** for performance). The supported targets are: `INTEL_GPU`, `NVIDIA_GPU`, `AMD_GPU` | | `CMAKE_PREFIX_PATH` | path | List of paths to check when searching for dependencies | | `CMAKE_INSTALL_PREFIX` | path | Specify the install location, used when invoking `ninja install` | diff --git a/cmake/Modules/SYCL.cmake b/cmake/Modules/SYCL.cmake index 71946ce45..2fe4d16b8 100644 --- a/cmake/Modules/SYCL.cmake +++ b/cmake/Modules/SYCL.cmake @@ -25,12 +25,12 @@ include(CheckCXXCompilerFlag) include(ConfigurePORTBLAS) -# find_package(AdaptiveCpp) requires HIPSYCL_TARGETS to be set, so set it to a default value before find_package(AdaptiveCpp) -if(SYCL_COMPILER MATCHES "adaptivecpp" AND NOT HIPSYCL_TARGETS AND NOT ENV{HIPSYCL_TARGETS}) - message(STATUS "Using `omp` as HIPSYCL_TARGETS") - set(HIPSYCL_TARGETS "omp") +# find_package(AdaptiveCpp) requires ACPP_TARGETS to be set, so set it to a default value before find_package(AdaptiveCpp) +if(SYCL_COMPILER MATCHES "adaptivecpp" AND NOT ACPP_TARGETS AND NOT ENV{ACPP_TARGETS}) + message(STATUS "Using `omp` as ACPP_TARGETS") + set(ACPP_TARGETS "omp") else() - message(STATUS "Using ${HIPSYCL_TARGETS} as HIPSYCL_TARGETS") + message(STATUS "Using ${ACPP_TARGETS} as ACPP_TARGETS") endif() check_cxx_compiler_flag("--acpp-targets" has_acpp) diff --git a/include/operations/blas_constants.h b/include/operations/blas_constants.h index 48e51a1bc..7d128dfbe 100644 --- a/include/operations/blas_constants.h +++ b/include/operations/blas_constants.h @@ -265,7 +265,7 @@ struct constant_pair { } // namespace blas -#ifndef __HIPSYCL__ +#ifndef __ADAPTIVECPP__ template struct cl::sycl::is_device_copyable> : std::true_type {}; diff --git a/include/sb_handle/portblas_handle.h b/include/sb_handle/portblas_handle.h index 6fd6faf98..836b37c61 100644 --- a/include/sb_handle/portblas_handle.h +++ b/include/sb_handle/portblas_handle.h @@ -49,13 +49,17 @@ class SB_Handle { public: using event_t = std::vector; inline SB_Handle(queue_t q) - : tempMemPool_(nullptr), + : +#ifndef __ADAPTIVECPP__ + tempMemPool_(nullptr), +#endif q_(q), workGroupSize_(helper::get_work_group_size(q)), localMemorySupport_(helper::has_local_memory(q)), - computeUnits_(helper::get_num_compute_units(q)) {} + computeUnits_(helper::get_num_compute_units(q)) { + } -#ifndef __HIPSYCL__ +#ifndef __ADAPTIVECPP__ inline SB_Handle(Temp_Mem_Pool* tmp) : tempMemPool_(tmp), q_(tmp->get_queue()), @@ -193,7 +197,9 @@ class SB_Handle { const size_t workGroupSize_; const bool localMemorySupport_; const size_t computeUnits_; +#ifndef __ADAPTIVECPP__ Temp_Mem_Pool* tempMemPool_; +#endif }; } // namespace blas diff --git a/include/sb_handle/temp_memory_pool.h b/include/sb_handle/temp_memory_pool.h index 8fb4831a7..836fb98c3 100644 --- a/include/sb_handle/temp_memory_pool.h +++ b/include/sb_handle/temp_memory_pool.h @@ -25,7 +25,7 @@ #ifndef TEMP_MEMORY_POOL_H #define TEMP_MEMORY_POOL_H -#ifndef __HIPSYCL__ +#ifndef __ADAPTIVECPP__ #include #include @@ -114,14 +114,6 @@ class Temp_Mem_Pool { }; } // namespace blas -#else - -namespace blas { -// Empty class to serve as a temporary Placeholder asthe feature -// is not fully supported when using AdaptiveCpp -class Temp_Mem_Pool; -} // namespace blas - -#endif // __HIPSYCL__ +#endif // __ADAPTIVECPP__ #endif diff --git a/src/interface/blas1/backend/default_cpu.hpp b/src/interface/blas1/backend/default_cpu.hpp index b897a2ba5..ac2fe764d 100644 --- a/src/interface/blas1/backend/default_cpu.hpp +++ b/src/interface/blas1/backend/default_cpu.hpp @@ -49,7 +49,7 @@ template ( @@ -85,7 +85,7 @@ template ( diff --git a/src/interface/blas1_interface.hpp b/src/interface/blas1_interface.hpp index 2c474d235..df914b71b 100644 --- a/src/interface/blas1_interface.hpp +++ b/src/interface/blas1_interface.hpp @@ -151,7 +151,7 @@ typename sb_handle_t::event_t _sdsdot( sb_handle_t &sb_handle, index_t _N, float sb, container_0_t _vx, increment_t _incx, container_1_t _vy, increment_t _incy, container_2_t _rs, const typename sb_handle_t::event_t &_dependencies) { -#ifndef __HIPSYCL__ +#ifndef __ADAPTIVECPP__ if (!_N) { using element_t = typename ValueType::type; sb_handle.wait(_dependencies); @@ -190,7 +190,7 @@ typename sb_handle_t::event_t _asum( sb_handle_t &sb_handle, index_t _N, container_0_t _vx, increment_t _incx, container_1_t _rs, const typename sb_handle_t::event_t &_dependencies) { // keep compatibility with older sycl versions -#if SYCL_LANGUAGE_VERSION < 202000 || defined(__HIPSYCL__) +#if SYCL_LANGUAGE_VERSION < 202000 || defined(__ADAPTIVECPP__) typename VectorViewType::type vx = make_vector_view(_vx, _incx, _N); auto rs = make_vector_view(_rs, static_cast(1), @@ -208,7 +208,7 @@ typename sb_handle_t::event_t _asum( #endif } -#if SYCL_LANGUAGE_VERSION >= 202000 && !defined(__HIPSYCL__) +#if SYCL_LANGUAGE_VERSION >= 202000 && !defined(__ADAPTIVECPP__) /*! _asum_impl. * @brief Internal implementation of the Absolute sum operator. * @@ -924,7 +924,7 @@ typename ValueType::type _dot( sb_handle_t &sb_handle, index_t _N, container_0_t _vx, increment_t _incx, container_1_t _vy, increment_t _incy, const typename sb_handle_t::event_t &_dependencies) { -#ifndef __HIPSYCL__ +#ifndef __ADAPTIVECPP__ constexpr bool is_usm = std::is_pointer::value; using element_t = typename ValueType::type; element_t res{0}; @@ -1074,7 +1074,7 @@ template ::type _asum( sb_handle_t &sb_handle, index_t _N, container_t _vx, increment_t _incx, const typename sb_handle_t::event_t &_dependencies) { -#ifndef __HIPSYCL__ +#ifndef __ADAPTIVECPP__ constexpr bool is_usm = std::is_pointer::value; using element_t = typename ValueType::type; auto res = std::vector(1, element_t(0)); @@ -1114,7 +1114,7 @@ template ::type _nrm2( sb_handle_t &sb_handle, index_t _N, container_t _vx, increment_t _incx, const typename sb_handle_t::event_t &_dependencies) { -#ifndef __HIPSYCL__ +#ifndef __ADAPTIVECPP__ constexpr bool is_usm = std::is_pointer::value; using element_t = typename ValueType::type; auto res = std::vector(1, element_t(0)); diff --git a/src/interface/blas2_interface.hpp b/src/interface/blas2_interface.hpp index ac07cf499..71dbee066 100644 --- a/src/interface/blas2_interface.hpp +++ b/src/interface/blas2_interface.hpp @@ -356,7 +356,7 @@ typename sb_handle_t::event_t _trsv_impl( sb_handle_t& sb_handle, index_t _N, container_t0 _mA, index_t _lda, container_t1 _vx, increment_t _incx, const typename sb_handle_t::event_t& _dependencies) { -#if (SYCL_LANGUAGE_VERSION < 202000) || (defined __HIPSYCL__) +#if (SYCL_LANGUAGE_VERSION < 202000) || (defined __ADAPTIVECPP__) throw std::runtime_error("trsv requires SYCL 2020"); #else static_assert(subgroup_size % subgroups == 0, @@ -747,7 +747,7 @@ typename sb_handle_t::event_t _tbsv_impl( sb_handle_t& sb_handle, index_t _N, index_t _K, container_t0 _mA, index_t _lda, container_t1 _vx, increment_t _incx, const typename sb_handle_t::event_t& _dependencies) { -#if (SYCL_LANGUAGE_VERSION < 202000) || (defined __HIPSYCL__) +#if (SYCL_LANGUAGE_VERSION < 202000) || (defined __ADAPTIVECPP__) throw std::runtime_error("tbsv requires SYCL 2020"); #else static_assert(subgroup_size % subgroups == 0, @@ -810,7 +810,7 @@ template ::eval( using element_t = typename ResolveReturnType::type::value_t::value_t; -#ifndef __HIPSYCL__ +#ifndef __ADAPTIVECPP__ // reduction within the sub_group for (index_t i = sg_local_range >> 1; i > 0; i >>= 1) { if (sg_local_id < i) { diff --git a/src/operations/blas2/txsv.hpp b/src/operations/blas2/txsv.hpp index 55e1f7fb9..a974253f2 100644 --- a/src/operations/blas2/txsv.hpp +++ b/src/operations/blas2/txsv.hpp @@ -100,7 +100,7 @@ Txsv::eval(local_memory_t local_mem, cl::sycl::nd_item<1> ndItem) { value_t ret = 0; -#if (SYCL_LANGUAGE_VERSION >= 202000) && !(defined __HIPSYCL__) +#if (SYCL_LANGUAGE_VERSION >= 202000) && !(defined __ADAPTIVECPP__) constexpr bool is_forward = (is_upper && is_transposed) || (!is_upper && !is_transposed); diff --git a/src/sb_handle/portblas_handle.hpp b/src/sb_handle/portblas_handle.hpp index 22901719f..f158de242 100644 --- a/src/sb_handle/portblas_handle.hpp +++ b/src/sb_handle/portblas_handle.hpp @@ -44,7 +44,7 @@ typename std::enable_if< alloc == helper::AllocType::buffer, typename helper::AllocHelper::type>::type SB_Handle::acquire_temp_mem(size_t size) { -#ifndef __HIPSYCL__ +#ifndef __ADAPTIVECPP__ if (tempMemPool_ != nullptr) return tempMemPool_->acquire_buff_mem(size); else @@ -60,7 +60,7 @@ typename std::enable_if< typename SB_Handle::event_t>::type SB_Handle::release_temp_mem(const typename SB_Handle::event_t& dependencies, const container_t& mem) { -#ifndef __HIPSYCL__ +#ifndef __ADAPTIVECPP__ if (tempMemPool_ != nullptr) return tempMemPool_->release_buff_mem(dependencies, mem); else diff --git a/src/sb_handle/temp_memory_pool.hpp b/src/sb_handle/temp_memory_pool.hpp index 2fdf7c4ed..1d57b0c6f 100644 --- a/src/sb_handle/temp_memory_pool.hpp +++ b/src/sb_handle/temp_memory_pool.hpp @@ -1,6 +1,6 @@ #ifndef TEMP_MEMORY_POOL_HPP #define TEMP_MEMORY_POOL_HPP -#ifndef __HIPSYCL__ +#ifndef __ADAPTIVECPP__ #include "portblas_helper.h" namespace blas { @@ -115,5 +115,5 @@ typename Temp_Mem_Pool::event_t Temp_Mem_Pool::release_usm_mem( } } #endif // SB_ENABLE_USM -#endif // __HIPSYCL__ +#endif // __ADAPTIVECPP__ #endif