diff --git a/include/oneapi/dpl/pstl/hetero/dpcpp/sycl_iterator.h b/include/oneapi/dpl/pstl/hetero/dpcpp/sycl_iterator.h index b238133ace..9109685f69 100644 --- a/include/oneapi/dpl/pstl/hetero/dpcpp/sycl_iterator.h +++ b/include/oneapi/dpl/pstl/hetero/dpcpp/sycl_iterator.h @@ -131,23 +131,31 @@ struct sycl_iterator } }; -// mode converter when property::noinit present -template -struct _ModeConverter +// map access_mode tag to access_mode value +// TODO: consider removing the logic for discard_read_write and discard_write which are deprecated in SYCL 2020 +template +struct __access_mode_resolver { - static constexpr access_mode __value = __mode; }; -template <> -struct _ModeConverter +template +struct __access_mode_resolver, _NoInitT> { - static constexpr access_mode __value = access_mode::discard_read_write; + static constexpr access_mode __mode = access_mode::read; }; -template <> -struct _ModeConverter +template +struct __access_mode_resolver, _NoInitT> { - static constexpr access_mode __value = access_mode::discard_write; + static constexpr access_mode __value = + std::is_same_v<_NoInitT, void> ? access_mode::write : access_mode::discard_write; +}; + +template +struct __access_mode_resolver, _NoInitT> +{ + static constexpr access_mode __value = + std::is_same_v<_NoInitT, void> ? access_mode::read_write : access_mode::discard_read_write; }; template ::value_type>> @@ -192,17 +200,19 @@ __internal::sycl_iterator end(sycl::buffe } // begin -template -__internal::sycl_iterator begin(sycl::buffer buf, sycl::mode_tag_t) +template +__internal::sycl_iterator<__internal::__access_mode_resolver::__value, T, Allocator> +begin(sycl::buffer buf, ModeTagT) { - return __internal::sycl_iterator{buf, 0}; + return __internal::sycl_iterator<__internal::__access_mode_resolver::__value, T, Allocator>{buf, 0}; } -template -__internal::sycl_iterator<__internal::_ModeConverter::__value, T, Allocator> - begin(sycl::buffer buf, sycl::mode_tag_t, __dpl_sycl::__no_init) +template +__internal::sycl_iterator<__internal::__access_mode_resolver::__value, T, Allocator> +begin(sycl::buffer buf, ModeTagT, __dpl_sycl::__no_init) { - return __internal::sycl_iterator<__internal::_ModeConverter::__value, T, Allocator>{buf, 0}; + return __internal::sycl_iterator<__internal::__access_mode_resolver::__value, T, + Allocator>{buf, 0}; } template @@ -213,18 +223,20 @@ __internal::sycl_iterator } // end -template -__internal::sycl_iterator end(sycl::buffer buf, sycl::mode_tag_t) +template +__internal::sycl_iterator<__internal::__access_mode_resolver::__value, T, Allocator> +end(sycl::buffer buf, ModeTagT) { - return __internal::sycl_iterator{buf, __dpl_sycl::__get_buffer_size(buf)}; + return __internal::sycl_iterator<__internal::__access_mode_resolver::__value, T, Allocator>{ + buf, __dpl_sycl::__get_buffer_size(buf)}; } -template -__internal::sycl_iterator<__internal::_ModeConverter::__value, T, Allocator> - end(sycl::buffer buf, sycl::mode_tag_t, __dpl_sycl::__no_init) +template +__internal::sycl_iterator<__internal::__access_mode_resolver::__value, T, Allocator> +end(sycl::buffer buf, ModeTagT, __dpl_sycl::__no_init) { - return __internal::sycl_iterator<__internal::_ModeConverter::__value, T, Allocator>{ - buf, __dpl_sycl::__get_buffer_size(buf)}; + return __internal::sycl_iterator<__internal::__access_mode_resolver::__value, T, + Allocator>{buf, __dpl_sycl::__get_buffer_size(buf)}; } template