Addressed more PR reviews

codeplaysoftware · Jan 29, 2024 · 199717f · 199717f
1 parent c8d3044
commit 199717f
Show file tree

Hide file tree

Showing 12 changed files with 38 additions and 40 deletions.
diff --git 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
@@ -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
@@ -265,7 +265,7 @@ struct constant_pair {
 
 }  // namespace blas
 
-#ifndef __HIPSYCL__
+#ifndef __ADAPTIVECPP__
 template <typename ind_t, typename val_t>
 struct cl::sycl::is_device_copyable<blas::IndexValueTuple<ind_t, val_t>>
     : std::true_type {};

diff --git 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<cl::sycl::event>;
   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
@@ -25,7 +25,7 @@
 #ifndef TEMP_MEMORY_POOL_H
 #define TEMP_MEMORY_POOL_H
 
-#ifndef __HIPSYCL__
+#ifndef __ADAPTIVECPP__
 #include <map>
 #include <mutex>
 
@@ -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
@@ -49,7 +49,7 @@ template <typename sb_handle_t, typename container_0_t, typename container_1_t,
 typename sb_handle_t::event_t _iamax(
     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) {
-#ifndef __HIPSYCL__
+#ifndef __ADAPTIVECPP__
   constexpr int localSize = 128;
   if (_N < 8192) {
     return blas::internal::_iamax_iamin_impl<localSize, 0, true, true>(
@@ -85,7 +85,7 @@ template <typename sb_handle_t, typename container_0_t, typename container_1_t,
 typename sb_handle_t::event_t _iamin(
     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) {
-#ifndef __HIPSYCL__
+#ifndef __ADAPTIVECPP__
   constexpr int localSize = 128;
   if (_N < 8192) {
     return blas::internal::_iamax_iamin_impl<localSize, 0, false, true>(

diff --git 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<container_2_t>::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<container_0_t, index_t, increment_t>::type vx =
       make_vector_view(_vx, _incx, _N);
   auto rs = make_vector_view(_rs, static_cast<increment_t>(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<container_0_t>::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<container_0_t>::value;
   using element_t = typename ValueType<container_0_t>::type;
   element_t res{0};
@@ -1074,7 +1074,7 @@ template <typename sb_handle_t, typename container_t, typename index_t,
 typename ValueType<container_t>::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<container_t>::value;
   using element_t = typename ValueType<container_t>::type;
   auto res = std::vector<element_t>(1, element_t(0));
@@ -1114,7 +1114,7 @@ template <typename sb_handle_t, typename container_t, typename index_t,
 typename ValueType<container_t>::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<container_t>::value;
   using element_t = typename ValueType<container_t>::type;
   auto res = std::vector<element_t>(1, element_t(0));

diff --git 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 <uint32_t subgroup_size, uint32_t subgroups, uplo_type uplo,
 typename sb_handle_t::event_t _tpsv_impl(
     sb_handle_t& sb_handle, index_t _N, container_t0 _mA, 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("tpsv requires SYCL 2020");
 #else
   static_assert(subgroup_size % subgroups == 0,

diff --git a/src/operations/blas1/IndexMaxMin.hpp b/src/operations/blas1/IndexMaxMin.hpp
@@ -94,7 +94,7 @@ PORTBLAS_INLINE void IndexMaxMin<is_max, is_step0, lhs_t, rhs_t>::eval(
   using element_t =
       typename ResolveReturnType<op, rhs_t>::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
@@ -100,7 +100,7 @@ Txsv<vector_t, matrix_t, sync_t, matrix_format, subgroup_size, subgroups,
      is_upper, is_transposed, is_unitdiag>::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
@@ -44,7 +44,7 @@ typename std::enable_if<
     alloc == helper::AllocType::buffer,
     typename helper::AllocHelper<value_t, alloc>::type>::type
 SB_Handle::acquire_temp_mem(size_t size) {
-#ifndef __HIPSYCL__
+#ifndef __ADAPTIVECPP__
   if (tempMemPool_ != nullptr)
     return tempMemPool_->acquire_buff_mem<value_t>(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
@@ -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