diff --git a/include/alpaka/kernel/TaskKernelGenericSycl.hpp b/include/alpaka/kernel/TaskKernelGenericSycl.hpp index 3e83dbcbf569..64e89190ba79 100644 --- a/include/alpaka/kernel/TaskKernelGenericSycl.hpp +++ b/include/alpaka/kernel/TaskKernelGenericSycl.hpp @@ -114,8 +114,7 @@ namespace alpaka::detail template inline auto require( sycl::handler& cgh, - experimental::Accessor::value, TAccessModes>, TElem, TIdx, TDim, TAccessModes> - acc, + Accessor::value, TAccessModes>, TElem, TIdx, TDim, TAccessModes> acc, special) { cgh.require(acc.m_accessor); diff --git a/include/alpaka/mem/buf/sycl/Accessor.hpp b/include/alpaka/mem/buf/sycl/Accessor.hpp index d9ff73e56c9b..c7a7e92a0bd9 100644 --- a/include/alpaka/mem/buf/sycl/Accessor.hpp +++ b/include/alpaka/mem/buf/sycl/Accessor.hpp @@ -20,14 +20,24 @@ namespace alpaka namespace detail { + //! Access tag type indicating read-only access. + struct ReadAccess + { + }; + + //! Access tag type indicating write-only access. + struct WriteAccess + { + }; + template inline constexpr auto sycl_access_mode = sycl::access_mode::read_write; template<> - inline constexpr auto sycl_access_mode = sycl::access_mode::read; + inline constexpr auto sycl_access_mode = sycl::access_mode::read; template<> - inline constexpr auto sycl_access_mode = sycl::access_mode::write; + inline constexpr auto sycl_access_mode = sycl::access_mode::write; template using SyclAccessor = sycl::accessor< @@ -38,15 +48,25 @@ namespace alpaka sycl::access::placeholder::true_t>; } // namespace detail + //! An accessor is an abstraction for accessing memory objects such as views and buffers. + //! @tparam TMemoryHandle A handle to a memory object. + //! @tparam TElem The type of the element stored by the memory object. Values and references to this type are + //! returned on access. + //! @tparam TBufferIdx The integral type used for indexing and index computations. + //! @tparam TDim The dimensionality of the accessed data. + //! @tparam TAccessModes Either a single access tag type or a `std::tuple` containing multiple access tag + //! types. + template + struct Accessor; + template - struct experimental:: - Accessor, TElem, TIdx, std::size_t{1}, TAccessModes> + struct Accessor, TElem, TIdx, std::size_t{1}, TAccessModes> { static constexpr auto sycl_access_mode = detail::sycl_access_mode; using SyclAccessor = detail::SyclAccessor; using VecType = Vec, TIdx>; using ReturnType = std::conditional_t< - std::is_same_v, + std::is_same_v, typename SyclAccessor::const_reference, typename SyclAccessor::reference>; @@ -77,14 +97,13 @@ namespace alpaka }; template - struct experimental:: - Accessor::value, TAccessModes>, TElem, TIdx, TDim, TAccessModes> + struct Accessor::value, TAccessModes>, TElem, TIdx, TDim, TAccessModes> { static constexpr auto sycl_access_mode = detail::sycl_access_mode; using SyclAccessor = detail::SyclAccessor::value, TAccessModes>; using VecType = Vec, TIdx>; using ReturnType = std::conditional_t< - std::is_same_v, + std::is_same_v, typename SyclAccessor::const_reference, typename SyclAccessor::reference>;