|
DPC++ Runtime
Runtime libraries for oneAPI DPC++
|
|
|
DPC++ Runtime
Runtime libraries for oneAPI DPC++
|
|
#include <sycl/ext/intel/esimd/common.hpp>#include <sycl/ext/intel/esimd/detail/memory_intrin.hpp>#include <sycl/ext/intel/esimd/detail/types.hpp>#include <sycl/ext/intel/esimd/detail/util.hpp>#include <sycl/ext/intel/esimd/memory_properties.hpp>#include <sycl/ext/intel/esimd/simd.hpp>#include <sycl/ext/intel/esimd/simd_view.hpp>#include <sycl/half_type.hpp>#include <algorithm>#include <cstdint>
Go to the source code of this file.
Namespaces | |
| sycl | |
| sycl::_V1 | |
| sycl::_V1::ext | |
| sycl::_V1::ext::intel | |
| sycl::_V1::ext::intel::esimd | |
| sycl::_V1::ext::intel::esimd::detail | |
Macros | |
| #define | __ESIMD_FP_ATOMIC_OP_TYPE_CHECK(T) |
Typedefs | |
| using | sycl::_V1::ext::intel::esimd::detail::DeviceAccessorOffsetT = uint32_t |
Enumerations | |
| enum | sycl::_V1::ext::intel::esimd::fence_mask : uint8_t { sycl::_V1::ext::intel::esimd::global_coherent_fence = 0x1 , sycl::_V1::ext::intel::esimd::l2_flush_instructions = 0x2 , sycl::_V1::ext::intel::esimd::__SYCL_DEPRECATED , sycl::_V1::ext::intel::esimd::l2_flush_texture_data = 0x4 , sycl::_V1::ext::intel::esimd::__SYCL_DEPRECATED , sycl::_V1::ext::intel::esimd::l2_flush_constant_data = 0x8 , sycl::_V1::ext::intel::esimd::__SYCL_DEPRECATED , sycl::_V1::ext::intel::esimd::l2_flush_rw_data = 0x10 , sycl::_V1::ext::intel::esimd::__SYCL_DEPRECATED , sycl::_V1::ext::intel::esimd::local_barrier = 0x20 , sycl::_V1::ext::intel::esimd::l1_flush_ro_data = 0x40 , sycl::_V1::ext::intel::esimd::__SYCL_DEPRECATED } |
| Represetns a bit mask to control behavior of esimd::fence. More... | |
Functions | |
| template<typename AccessorTy > | |
| __ESIMD_API SurfaceIndex | sycl::_V1::ext::intel::esimd::get_surface_index (AccessorTy acc) |
| Get surface index corresponding to a SYCL accessor. More... | |
| template<typename RT , typename T , int N> | |
| ESIMD_INLINE simd< RT, N > | sycl::_V1::ext::intel::esimd::detail::lsc_format_input (simd< T, N > Vals) |
| template<typename T , typename T1 , int N> | |
| ESIMD_INLINE simd< T, N > | sycl::_V1::ext::intel::esimd::detail::lsc_format_ret (simd< T1, N > Vals) |
| template<typename PropertyListT , cache_level Level> | |
| constexpr cache_hint | sycl::_V1::ext::intel::esimd::detail::getCacheHintForIntrin () |
| Extracts a cache hint with the given 'Level' to pass it to ESIMD/GENX intrinsics. More... | |
| template<typename T , int NElts, lsc_data_size DS, typename PropertyListT , int N, typename OffsetT > | |
| __ESIMD_API simd< T, N *NElts > | sycl::_V1::ext::intel::esimd::detail::gather_impl (const T *p, simd< OffsetT, N > offsets, simd_mask< N > pred, simd< T, N *NElts > pass_thru) |
| USM pointer gather. More... | |
| template<typename T , int NElts, lsc_data_size DS, typename PropertyListT , int N, typename Toffset > | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::detail::scatter_impl (T *p, simd< Toffset, N > offsets, simd< T, N *NElts > vals, simd_mask< N > pred) |
| USM pointer scatter. More... | |
| constexpr bool | sycl::_V1::ext::intel::esimd::detail::isMaskedGatherScatterLLVMAvailable () |
| template<typename T , int N, int VS, typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (const T *p, simd< OffsetT, N/VS > byte_offsets, simd_mask< N/VS > mask, simd< T, N > pass_thru, PropertyListT props={}) |
| template <typename T, int N, int VS, typename OffsetT, typename PropertyListT = empty_properties_t> simd<T, N> gather(const T *p, simd<OffsetT, N / VS> byte_offsets, simd_mask<N / VS> mask, simd<T, N> pass_thru, PropertyListT props = {}); // (usm-ga-1) simd<T, N> gather(const T *p, simd<OffsetT, N / VS> byte_offsets, simd_mask<N / VS> mask, PropertyListT props = {}); // (usm-ga-2) simd<T, N> gather(const T *p, simd<OffsetT, N / VS> byte_offsets, PropertyListT props = {}); // (usm-ga-3) More... | |
| template<typename T , int N, int VS, typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (const T *p, simd< OffsetT, N/VS > byte_offsets, simd_mask< N/VS > mask, PropertyListT props={}) |
template <typename T, int N, int VS, typename OffsetT, typename PropertyListT = empty_properties_t> simd<T, N> gather(const T *p, simd<OffsetT, N / VS> byte_offsets, simd_mask<N / VS> mask, PropertyListT props = {}); // (usm-ga-2) Loads ("gathers") elements of the type 'T' from memory locations addressed by the base pointer p and byte offsets byte_offsets, and returns the loaded elements. More... | |
| template<typename T , int N, int VS, typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (const T *p, simd< OffsetT, N/VS > byte_offsets, PropertyListT props={}) |
template <typename T, int N, int VS, typename OffsetT, typename PropertyListT = empty_properties_t> simd<T, N> gather(const T *p, simd<OffsetT, N / VS> byte_offsets, PropertyListT props = {}); // (usm-ga-3) Loads ("gathers") elements of the type 'T' from memory locations addressed by the base pointer p and byte offsets byte_offsets, and returns the loaded elements. More... | |
| template<typename T , int N, typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (const T *p, simd< OffsetT, N > byte_offsets, simd_mask< N > mask, simd< T, N > pass_thru, PropertyListT props={}) |
template <typename T, int N, typename OffsetT, typename PropertyListT = empty_properties_t> simd<T, N> gather(const T *p, simd<OffsetT, N> byte_offsets, simd_mask<N> mask, simd<T, N> pass_thru, PropertyListT props = {}); // (usm-ga-4) Loads ("gathers") elements of the type 'T' from memory locations addressed by the base pointer p and byte offsets byte_offsets, and returns the loaded elements. More... | |
| template<typename T , int N, typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (const T *p, simd< OffsetT, N > byte_offsets, simd_mask< N > mask, PropertyListT props={}) |
template <typename T, int N, typename OffsetT, typename PropertyListT = empty_properties_t> simd<T, N> gather(const T *p, simd<OffsetT, N> byte_offsets, simd_mask<N> mask, PropertyListT props = {});// (usm-ga-5) Loads ("gathers") elements of the type 'T' from memory locations addressed by the base pointer p and byte offsets byte_offsets, and returns the loaded elements. More... | |
| template<typename T , int N, typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (const T *p, simd< OffsetT, N > byte_offsets, PropertyListT props={}) |
template <typename T, int N, typename OffsetT, typename PropertyListT = empty_properties_t> simd<T, N> gather(const T *p, simd<OffsetT, N> byte_offsets, PropertyListT props = {}); // (usm-ga-6) Loads ("gathers") elements of the type 'T' from memory locations addressed by the base pointer p and byte offsets byte_offsets, and returns the loaded elements. More... | |
| template<typename T , int N, int VS = 1, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_simd_view_type_v< OffsetSimdViewT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (const T *p, OffsetSimdViewT byte_offsets, simd_mask< N/VS > mask, simd< T, N > pass_thru, PropertyListT props={}) |
template <typename T, int N, int VS = 1, typename OffsetSimdViewT, typename PropertyListT = empty_props_t> simd <T, N> gather(const T *p, OffsetSimdViewT byte_offsets, simd_mask<N / VS> mask, simd<T, N> pass_thru, PropertyListT props = {}); // (usm-ga-7) Loads ("gathers") elements of the type 'T' from memory locations addressed by the base pointer p and byte offsets byte_offsets, and returns the loaded elements. More... | |
| template<typename T , int N, int VS = 1, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_simd_view_type_v< OffsetSimdViewT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (const T *p, OffsetSimdViewT byte_offsets, simd_mask< N/VS > mask, PropertyListT props={}) |
simd <T, N> gather(const T *p, OffsetSimdViewT byte_offsets, simd_mask<N / VS> mask, PropertyListT props = {}); // (usm-ga-8) Loads ("gathers") elements of the type 'T' from memory locations addressed by the base pointer p and byte offsets byte_offsets, and returns the loaded elements. More... | |
| template<typename T , int N, int VS = 1, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_simd_view_type_v< OffsetSimdViewT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (const T *p, OffsetSimdViewT byte_offsets, PropertyListT props={}) |
simd <T, N> gather(const T *p, OffsetSimdViewT byte_offsets, PropertyListT props = {}); // (usm-ga-9) Loads ("gathers") elements of the type 'T' from memory locations addressed by the base pointer p and byte offsets byte_offsets, and returns the loaded elements. More... | |
| template<typename Tx , int N, typename Toffset > | |
| __ESIMD_API std::enable_if_t< std::is_integral_v< Toffset >, simd< Tx, N > > | sycl::_V1::ext::intel::esimd::gather (const Tx *p, Toffset offset, simd_mask< N > mask=1) |
A variation of gather API with offsets represented as scalar. More... | |
| template<typename T , int N, int VS = 1, typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::scatter (T *p, simd< OffsetT, N/VS > byte_offsets, simd< T, N > vals, simd_mask< N/VS > mask, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename OffsetT, typename PropertyListT = empty_properties_t> void scatter(T *p, simd<OffsetT, N / VS> byte_offsets, simd<T, N> vals, simd_mask<N / VS> mask, PropertyListT props = {}); // (usm-sc-1) More... | |
| template<typename T , int N, int VS = 1, typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::scatter (T *p, simd< OffsetT, N/VS > byte_offsets, simd< T, N > vals, PropertyListT props={}) |
| Writes ("scatters") elements of the input vector to different memory locations. More... | |
| template<typename T , int N, int VS = 1, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::scatter (T *p, OffsetSimdViewT byte_offsets, simd< T, N > vals, simd_mask< N/VS > mask, PropertyListT props={}) |
| Writes ("scatters") elements of the input vector to different memory locations. More... | |
| template<typename T , int N, int VS = 1, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::scatter (T *p, OffsetSimdViewT byte_offsets, simd< T, N > vals, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename OffsetSimdViewT, typename PropertyListT = empty_properties_t> void scatter(T *p, OffsetSimdViewT byte_offsets, simd<T, N> vals, PropertyListT props = {}); // (usm-sc-4) More... | |
| template<typename Tx , int N, typename Toffset > | |
| __ESIMD_API std::enable_if_t< std::is_integral_v< Toffset > &&N==1 > | sycl::_V1::ext::intel::esimd::scatter (Tx *p, Toffset offset, simd< Tx, N > vals, simd_mask< N > mask=1) |
A variation of scatter API with offsets represented as scalar. More... | |
| template<typename T , int NElts, typename PropertyListT > | |
| __ESIMD_API std::enable_if_t< is_property_list_v< PropertyListT >, simd< T, NElts > > | sycl::_V1::ext::intel::esimd::detail::block_load_impl (const T *p, simd_mask< 1 > pred, simd< T, NElts > pass_thru) |
| USM pointer transposed gather with 1 channel. More... | |
| template<typename T , int NElts, typename PropertyListT , typename AccessorT > | |
| __ESIMD_API std::enable_if_t< detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&is_property_list_v< PropertyListT >, simd< T, NElts > > | sycl::_V1::ext::intel::esimd::detail::block_load_impl (AccessorT acc, DeviceAccessorOffsetT offset, simd_mask< 1 > pred) |
| Accessor-based transposed gather with 1 channel. More... | |
| template<typename T , int NElts, typename PropertyListT , typename AccessorT > | |
| __ESIMD_API std::enable_if_t< detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&is_property_list_v< PropertyListT >, simd< T, NElts > > | sycl::_V1::ext::intel::esimd::detail::block_load_impl (AccessorT acc, DeviceAccessorOffsetT offset, simd_mask< 1 > pred, simd< T, NElts > pass_thru) |
| Accessor-based transposed gather with 1 channel. More... | |
| template<typename T , int NElts, typename PropertyListT > | |
| __ESIMD_API std::enable_if_t< detail::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::detail::block_store_impl (T *p, simd< T, NElts > vals, simd_mask< 1 > pred) |
| template<typename T , int NElts, typename PropertyListT , typename AccessorT > | |
| __ESIMD_API std::enable_if_t< detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_write > &&detail::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::detail::block_store_impl (AccessorT acc, DeviceAccessorOffsetT offset, simd< T, NElts > vals, simd_mask< 1 > pred) |
| template<typename Tx , int N, typename Flags = overaligned_tag<detail::OperandSize::OWORD>> | |
| __ESIMD_API std::enable_if_t< is_simd_flag_type_v< Flags > > | sycl::_V1::ext::intel::esimd::block_store (Tx *addr, simd< Tx, N > vals, Flags) |
Stores elements of the vector vals to a contiguous block of memory at the given address addr. More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (const T *ptr, PropertyListT props={}) |
| Each of the following block load functions loads a contiguous memory block from the address referenced by the USM pointer 'ptr', or from 'ptr + offset', where 'offset' is the offset in bytes (not in elements!). More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (const T *ptr, size_t byte_offset, PropertyListT props={}) |
simd<T, N> block_load(const T* ptr, size_t byte_offset, props={}); // (usm-bl-2) This function loads a contiguous memory block from address referenced by USM pointer ptr and the given byte_offset. More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (const T *ptr, simd_mask< 1 > pred, PropertyListT props={}) |
simd<T, N> block_load(const T* ptr, simd_mask<1> pred, props={}); // (usm-bl-3) This function loads a contiguous memory block from USM pointer ptr. More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (const T *ptr, size_t byte_offset, simd_mask< 1 > pred, PropertyListT props={}) |
simd<T, N> block_load(const T* ptr, size_t byte_offset, simd_mask<1> pred, props={}); // (usm-bl-4) This function loads a contiguous memory block from address referenced by USM pointer ptr and the given byte_offset. More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (const T *ptr, simd_mask< 1 > pred, simd< T, N > pass_thru, PropertyListT props={}) |
simd<T, N> block_load(const T* ptr, simd_mask<1> pred, simd<T, N> pass_thru, props={}); // (usm-bl-5) This function loads a contiguous memory block from USM pointer ptr. More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (const T *ptr, size_t byte_offset, simd_mask< 1 > pred, simd< T, N > pass_thru, PropertyListT props={}) |
simd<T, N> block_load(const T* ptr, size_t byte_offset, simd_mask<1> pred, simd<T, N> pass_thru, props={}); // (usm-bl-6) This function loads a contiguous memory block from address referenced by USM pointer ptr and the given byte_offset. More... | |
| template<typename Tx , int N, typename Flags = overaligned_tag<detail::OperandSize::OWORD>> | |
| __ESIMD_API std::enable_if_t< is_simd_flag_type_v< Flags >, simd< Tx, N > > | sycl::_V1::ext::intel::esimd::block_load (const Tx *addr, Flags) |
Loads a contiguous block of memory from the given memory address addr and returns the loaded data as a vector. More... | |
| template<typename Tx , int N, typename AccessorTy , typename Flags = vector_aligned_tag, typename = std::enable_if_t< is_simd_flag_type_v<Flags> && detail::is_device_accessor_with_v< AccessorTy, detail::accessor_mode_cap::can_read>>, class T = detail::__raw_t<Tx>> | |
| __ESIMD_API simd< Tx, N > | sycl::_V1::ext::intel::esimd::block_load (AccessorTy acc, detail::DeviceAccessorOffsetT byte_offset, Flags flags) |
Loads a contiguous block of memory from the given accessor acc and byte_offset and returns the loaded data as a vector. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (AccessorT acc, detail::DeviceAccessorOffsetT byte_offset, PropertyListT props={}) |
| Each of the following block load functions loads a contiguous memory block from the address referenced by accessor 'acc', or from 'acc + byte_offset', The parameter 'pred' is the one element predicate. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (AccessorT acc, PropertyListT={}) |
simd<T, N> block_load(AccessorT acc, props = {}); // (acc-bl-2) This function loads a contiguous memory block referenced by accessor acc and implied offset=0. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (AccessorT acc, detail::DeviceAccessorOffsetT byte_offset, simd_mask< 1 > pred, simd< T, N > pass_thru, PropertyListT={}) |
simd<T, N> block_load(AccessorT acc, OffsetT byte_offset, simd_mask<1> pred, simd<T, N> pass_thru, props = {}); // (acc-bl-3) This function loads a contiguous memory block referenced by accessor acc and the given byte_offset. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (AccessorT acc, detail::DeviceAccessorOffsetT byte_offset, simd_mask< 1 > pred, PropertyListT props={}) |
simd<T, N> block_load(AccessorT acc, OffsetT byte_offset, simd_mask<1> pred, props = {}); // (acc-bl-4) This function loads a contiguous memory block referenced by accessor acc and the given byte_offset. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (AccessorT acc, simd_mask< 1 > pred, simd< T, N > pass_thru, PropertyListT={}) |
simd<T, N> block_load(AccessorT acc, simd_mask<1> pred, simd<T, N> pass_thru, props = {}); // (acc-bl-5) This function loads a contiguous memory block referenced by accessor acc and implied offset=0. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (AccessorT acc, simd_mask< 1 > pred, PropertyListT={}) |
simd<T, N> block_load(AccessorT acc, simd_mask<1> pred, props = {}); // (acc-bl-6) This function loads a contiguous memory block referenced by accessor acc and implied offset=0. More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::block_store (T *ptr, simd< T, N > vals, PropertyListT={}) |
| Each of the following block store functions stores a contiguous memory block to the address referenced by the USM pointer 'ptr', or from 'ptr + offset', where 'offset' is the offset in bytes (not in elements!) with data specified by 'vals'. More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::block_store (T *ptr, size_t byte_offset, simd< T, N > vals, PropertyListT props={}) |
void block_store(T* ptr, size_t byte_offset, // (usm-bs-2) simd<T, N> vals, props={}); This function stores a contiguous memory block to USM pointer ptr and byte-offset byte_offset with data specified by vals. More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::block_store (T *ptr, simd< T, N > vals, simd_mask< 1 > pred, PropertyListT={}) |
void block_store(T* ptr, simd<T, N> vals, // (usm-bs-3) simd_mask<1> pred, props={}); This function stores a contiguous memory block to USM pointer ptr with data specified by vals. More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::block_store (T *ptr, size_t byte_offset, simd< T, N > vals, simd_mask< 1 > pred, PropertyListT props={}) |
void block_store(T* ptr, size_t byte_offset, // (usm-bs-4) simd<T, N> vals, simd_mask<1> pred, props={}); This function stores a contiguous memory block to USM pointer ptr and byte-offset byte_offset with data specified by vals. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_write > > | sycl::_V1::ext::intel::esimd::block_store (AccessorT acc, detail::DeviceAccessorOffsetT byte_offset, simd< T, N > vals, PropertyListT props={}) |
| Each of the following block_store functions stores the vector 'vals' to a contiguous memory block at the address referenced by accessor 'acc', or from 'acc + byte_offset', The parameter 'pred' is the one element predicate. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_write > > | sycl::_V1::ext::intel::esimd::block_store (AccessorT acc, simd< T, N > vals, PropertyListT props={}) |
void block_store(AccessorT acc, simd<T, N> vals, props = {}); // (acc-bs-2) This function stores a contiguous memory block to accessor acc with data specified by vals and implied offset=0. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_write > > | sycl::_V1::ext::intel::esimd::block_store (AccessorT acc, detail::DeviceAccessorOffsetT byte_offset, simd< T, N > vals, simd_mask< 1 > pred, PropertyListT props={}) |
void block_store(AccessorT acc, OffsetT byte_offset, // (acc-bs-3) simd<T, N> vals, simd_mask<1> pred, props = {}); This function stores a contiguous memory block to accessor acc and byte_offset with data specified by vals. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_write > > | sycl::_V1::ext::intel::esimd::block_store (AccessorT acc, simd< T, N > vals, simd_mask< 1 > pred, PropertyListT props={}) |
void block_store(AccessorT acc, simd<T, N> vals, // (acc-bs-4) simd_mask<1> pred, props = {}); This function stores a contiguous memory block to accessor acc with data specified by vals and implied offset=0. More... | |
| template<typename T , int N, typename AccessorT > | |
| __ESIMD_API std::enable_if_t< detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read >, simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, simd< detail::DeviceAccessorOffsetT, N > byte_offsets, detail::DeviceAccessorOffsetT glob_offset, simd_mask< N > mask=1) |
| Accessor-based gather. More... | |
| template<typename T , int N, typename AccessorT > | |
| __ESIMD_API std::enable_if_t< detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read >, simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, detail::DeviceAccessorOffsetT glob_offset) |
Loads and broadcasts the element located at acc and byte offset glob_offset to a vector and returns it as a simd object. More... | |
| template<typename T , int N, int VS, typename AccessorT , typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, simd< OffsetT, N/VS > byte_offsets, simd_mask< N/VS > mask, simd< T, N > pass_thru, PropertyListT props={}) |
| template <typename T, int N, int VS, typename AccessorT, typename OffsetT, typename PropertyListT = empty_properties_t> simd<T, N> gather(AccessorT acc, simd<OffsetT, N / VS> byte_offsets, simd_mask<N / VS> mask, simd<T, N> pass_thru, PropertyListT props = {}); // (acc-ga-1) simd<T, N> gather(AccessorT acc, simd<OffsetT, N / VS> byte_offsets, simd_mask<N / VS> mask, PropertyListT props = {}); // (acc-ga-2) simd<T, N> gather(AccessorT acc, simd<OffsetT, N / VS> byte_offsets, PropertyListT props = {}); // (acc-ga-3) More... | |
| template<typename T , int N, int VS, typename AccessorT , typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, simd< OffsetT, N/VS > byte_offsets, simd_mask< N/VS > mask, PropertyListT props={}) |
| template <typename T, int N, int VS, typename AccessorT, typename OffsetT, typename PropertyListT = empty_properties_t> simd<T, N> gather(AccessorT acc, simd<OffsetT, N / VS> byte_offsets, simd_mask<N / VS> mask, PropertyListT props = {}); // (acc-ga-2) Supported platforms: DG2, PVC in most cases. More... | |
| template<typename T , int N, int VS, typename AccessorT , typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, simd< OffsetT, N/VS > byte_offsets, PropertyListT props={}) |
| template <typename T, int N, int VS, typename AccessorT, typename OffsetT, typename PropertyListT = empty_properties_t> simd<T, N> gather(AccessorT acc, simd<OffsetT, N / VS> byte_offsets, PropertyListT props = {}); // (acc-ga-3) Supported platforms: DG2, PVC in most cases. More... | |
| template<typename T , int N, typename AccessorT , typename OffsetT , typename MaskT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&std::is_same_v< MaskT, simd_mask< N >> &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, simd< OffsetT, N > byte_offsets, MaskT mask, simd< T, N > pass_thru, PropertyListT props={}) |
| template <typename T, int N, typename AccessorT, typename OffsetT, typename PropertyListT = empty_properties_t> simd<T, N> gather(AccessorT acc, simd<OffsetT, N> byte_offsets, simd_mask<N> mask, simd<T, N> pass_thru, PropertyListT props = {}); // (acc-ga-4) This function is identical to (acc-ga-1) except that vector size is fixed to 1. More... | |
| template<typename T , int N, typename AccessorT , typename OffsetT , typename MaskT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&std::is_same_v< MaskT, simd_mask< N >> &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, simd< OffsetT, N > byte_offsets, MaskT mask, PropertyListT props={}) |
| template <typename T, int N, typename AccessorT, typename OffsetT, typename PropertyListT = empty_properties_t> simd<T, N> gather(AccessorT acc, simd<OffsetT, N> byte_offsets, simd_mask<N> mask, PropertyListT props // (acc-ga-5) This function is identical to (acc-ga-2) except that vector size is fixed to 1. More... | |
| template<typename T , int N, typename AccessorT , typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, simd< OffsetT, N > byte_offsets, PropertyListT props={}) |
| template <typename T, int N, typename AccessorT, typename OffsetT, typename PropertyListT = empty_properties_t> simd<T, N> gather(AccessorT acc, simd<OffsetT, N> byte_offsets, PropertyListT props = {}); // (acc-ga-6) This function is identical to (acc-ga-3) except that vector size is fixed to 1. More... | |
| template<typename T , int N, int VS = 1, typename AccessorT , typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, OffsetSimdViewT byte_offsets, simd_mask< N/VS > mask, simd< T, N > pass_thru, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename AccessorT, typename OffsetSimdViewT, More... | |
| template<typename T , int N, int VS = 1, typename AccessorT , typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, OffsetSimdViewT byte_offsets, simd_mask< N/VS > mask, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename AccessorT, typename OffsetSimdViewT, More... | |
| template<typename T , int N, int VS = 1, typename AccessorT , typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, OffsetSimdViewT byte_offsets, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename AccessorT, typename OffsetSimdViewT, More... | |
| template<typename T , int N, int VS = 1, typename AccessorTy , typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_device_accessor_with_v< AccessorTy, detail::accessor_mode_cap::can_write > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::scatter (AccessorTy acc, simd< OffsetT, N/VS > byte_offsets, simd< T, N > vals, simd_mask< N/VS > mask, PropertyListT props={}) |
| Accessor-based scatter. More... | |
| template<typename T , int N, int VS = 1, typename AccessorTy , typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_device_accessor_with_v< AccessorTy, detail::accessor_mode_cap::can_write > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::scatter (AccessorTy acc, simd< OffsetT, N/VS > byte_offsets, simd< T, N > vals, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename AccessorTy, typename OffsetT, typename PropertyListT = empty_properties_t> void scatter(AccessorTy acc, simd<OffsetT, N / VS> byte_offsets, simd<T, N> vals, PropertyListT props = {}); // (acc-sc-2) More... | |
| template<typename T , int N, int VS = 1, typename AccessorTy , typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_device_accessor_with_v< AccessorTy, detail::accessor_mode_cap::can_write > &&detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::scatter (AccessorTy acc, OffsetSimdViewT byte_offsets, simd< T, N > vals, simd_mask< N/VS > mask, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename AccessorTy, typename OffsetSimdViewT, typename PropertyListT = empty_properties_t> void scatter(AccessorTy acc, OffsetSimdViewT byte_offsets, simd<T, N> vals, simd_mask<N / VS> mask, PropertyListT props = {}); // (acc-sc-3) More... | |
| template<typename T , int N, int VS = 1, typename AccessorTy , typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_device_accessor_with_v< AccessorTy, detail::accessor_mode_cap::can_write > &&detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::scatter (AccessorTy acc, OffsetSimdViewT byte_offsets, simd< T, N > vals, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename AccessorTy, typename OffsetSimdViewT, typename PropertyListT = empty_properties_t> void scatter(AccessorTy acc, OffsetSimdViewT byte_offsets, simd<T, N> vals, PropertyListT props = {}); // (acc-sc-4) More... | |
| template<typename T , int N, typename AccessorTy > | |
| __ESIMD_API std::enable_if_t<(detail::isPowerOf2(N, 32)) &&detail::is_device_accessor_with_v< AccessorTy, detail::accessor_mode_cap::can_write > > | sycl::_V1::ext::intel::esimd::scatter (AccessorTy acc, simd< detail::DeviceAccessorOffsetT, N > offsets, simd< T, N > vals, detail::DeviceAccessorOffsetT glob_offset, simd_mask< N > mask=1) |
| Writes elements of a simd object into an accessor at given offsets. More... | |
| template<typename T , int N, typename AccessorTy > | |
| __ESIMD_API std::enable_if_t<(detail::isPowerOf2(N, 32)) &&detail::is_device_accessor_with_v< AccessorTy, detail::accessor_mode_cap::can_write > > | sycl::_V1::ext::intel::esimd::scatter (AccessorTy acc, detail::DeviceAccessorOffsetT glob_offset, simd< T, N > vals, simd_mask< N > mask=1) |
| template<typename T , typename AccessorTy > | |
| __ESIMD_API T | sycl::_V1::ext::intel::esimd::scalar_load (AccessorTy acc, detail::DeviceAccessorOffsetT offset) |
| Load a scalar value from an accessor. More... | |
| template<typename T , typename AccessorTy > | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::scalar_store (AccessorTy acc, detail::DeviceAccessorOffsetT offset, T val) |
| Store a scalar value into an accessor. More... | |
| template<rgba_channel_mask RGBAMask = rgba_channel_mask::ABGR, typename T , int N, typename Toffset > | |
| __ESIMD_API simd< T, N *get_num_channels_enabled(RGBAMask)> | sycl::_V1::ext::intel::esimd::gather_rgba (const T *p, simd< Toffset, N > offsets, simd_mask< N > mask=1) |
Gather and transpose pixels from given memory locations defined by the base pointer p and offsets. More... | |
| template<rgba_channel_mask RGBAMask = rgba_channel_mask::ABGR, typename T , int N, typename OffsetSimdViewT , typename RegionTy > | |
| __ESIMD_API std::enable_if_t< detail::is_simd_view_type_v< OffsetSimdViewT >, simd< T, N *get_num_channels_enabled(RGBAMask)> > | sycl::_V1::ext::intel::esimd::gather_rgba (const T *p, OffsetSimdViewT offsets, simd_mask< N > mask=1) |
A variation of gather_rgba API with offsets represented as simd_view object. More... | |
| template<rgba_channel_mask RGBAMask = rgba_channel_mask::ABGR, typename T , int N, typename Toffset > | |
| __ESIMD_API std::enable_if_t< std::is_integral_v< Toffset >, simd< T, N *get_num_channels_enabled(RGBAMask)> > | sycl::_V1::ext::intel::esimd::gather_rgba (const T *p, Toffset offset, simd_mask< N > mask=1) |
A variation of gather_rgba API with offsets represented as scalar. More... | |
| template<rgba_channel_mask M> | |
| static void | sycl::_V1::ext::intel::esimd::detail::validate_rgba_write_channel_mask () |
| template<rgba_channel_mask RGBAMask = rgba_channel_mask::ABGR, typename T , int N, typename Toffset > | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::scatter_rgba (T *p, simd< Toffset, N > offsets, simd< T, N *get_num_channels_enabled(RGBAMask)> vals, simd_mask< N > mask=1) |
Transpose and scatter pixels to given memory locations defined by the base pointer p and offsets. More... | |
| template<rgba_channel_mask RGBAMask = rgba_channel_mask::ABGR, typename T , int N, typename OffsetSimdViewT , typename RegionTy > | |
| __ESIMD_API std::enable_if_t< detail::is_simd_view_type_v< OffsetSimdViewT > > | sycl::_V1::ext::intel::esimd::scatter_rgba (T *p, OffsetSimdViewT offsets, simd< T, N *get_num_channels_enabled(RGBAMask)> vals, simd_mask< N > mask=1) |
A variation of scatter_rgba API with offsets represented as simd_view object. More... | |
| template<rgba_channel_mask RGBAMask = rgba_channel_mask::ABGR, typename T , int N, typename Toffset > | |
| __ESIMD_API std::enable_if_t< std::is_integral_v< Toffset > &&N==1 > | sycl::_V1::ext::intel::esimd::scatter_rgba (T *p, Toffset offset, simd< T, N *get_num_channels_enabled(RGBAMask)> vals, simd_mask< N > mask=1) |
A variation of scatter_rgba API with offsets represented as scalar. More... | |
| template<typename T , int N, rgba_channel_mask RGBAMask> | |
| __ESIMD_API std::enable_if_t<(N==8||N==16||N==32) &&sizeof(T)==4 > | sycl::_V1::ext::intel::esimd::scatter_rgba (T *p, simd< uint32_t, N > offsets, simd< T, N *get_num_channels_enabled(RGBAMask)> vals, simd_mask< N > mask=1) |
| template<rgba_channel_mask RGBAMask = rgba_channel_mask::ABGR, typename AccessorT , int N, typename T = typename AccessorT::value_type> | |
| __ESIMD_API std::enable_if_t<((N==8||N==16||N==32) &&sizeof(T)==4 &&detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read >), simd< T, N *get_num_channels_enabled(RGBAMask)> > | sycl::_V1::ext::intel::esimd::gather_rgba (AccessorT acc, simd< detail::DeviceAccessorOffsetT, N > offsets, detail::DeviceAccessorOffsetT global_offset=0, simd_mask< N > mask=1) |
Gather and transpose pixels from the given memory locations defined by the base specified by acc, the global offset global_offset and a vector of offsets offsets. More... | |
| template<rgba_channel_mask RGBAMask = rgba_channel_mask::ABGR, typename AccessorT , int N, typename T = typename AccessorT::value_type> | |
| __ESIMD_API std::enable_if_t<(N==8||N==16||N==32) &&sizeof(T)==4 &&detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_write > > | sycl::_V1::ext::intel::esimd::scatter_rgba (AccessorT acc, simd< detail::DeviceAccessorOffsetT, N > offsets, simd< T, N *get_num_channels_enabled(RGBAMask)> vals, detail::DeviceAccessorOffsetT global_offset=0, simd_mask< N > mask=1) |
Gather data from the memory addressed by accessor acc, offset common for all loaded elements global_offset and per-element offsets offsets, and return it as simd vector. More... | |
| template<sycl::ext::intel::esimd::atomic_op Op, typename T , int N, unsigned NumSrc, bool IsLSC = false> | |
| constexpr void | sycl::_V1::ext::intel::esimd::detail::check_atomic () |
| Check the legality of an atomic call in terms of size and type. More... | |
| template<uint32_t SLMSize> | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::slm_init () |
| Declare per-work-group slm size. More... | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::slm_init (uint32_t size) |
| Declare per-work-group slm size. More... | |
| template<typename T , int N, int VS, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::slm_gather (simd< uint32_t, N/VS > byte_offsets, simd_mask< N/VS > mask, simd< T, N > pass_thru, PropertyListT props={}) |
| template <typename T, int N, int VS, typename PropertyListT = empty_properties_t> simd<T, N> slm_gather(simd<uint32_t, N / VS> byte_offsets, simd_mask<N / VS> mask, simd<T, N> pass_thru, PropertyListT props = {}); // (slm-ga-1) simd<T, N> slm_gather(simd<uint32_t, N / VS> byte_offsets, simd_mask<N / VS> mask, PropertyListT props = {}); // (slm-ga-2) simd<T, N> slm_gather(simd<uint32_t, N / VS> byte_offsets, PropertyListT props = {}); // (slm-ga-3) More... | |
| template<typename T , int N, int VS, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::slm_gather (simd< uint32_t, N/VS > byte_offsets, simd_mask< N/VS > mask, PropertyListT props={}) |
template <typename T, int N, int VS, typename PropertyListT = empty_properties_t> simd<T, N> slm_gather(simd<uint32_t, N / VS> byte_offsets, simd_mask<N / VS> mask, PropertyListT props = {}); // (slm-ga-2) Loads ("gathers") elements of the type 'T' from Shared Local Memory locations addressed by byte offsets byte_offsets, and returns the loaded elements. More... | |
| template<typename T , int N, int VS, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::slm_gather (simd< uint32_t, N/VS > byte_offsets, PropertyListT props={}) |
template <typename T, int N, int VS, typename PropertyListT = empty_properties_t> simd<T, N> slm_gather(simd<uint32_t, N / VS> byte_offsets, PropertyListT props = {}); // (slm-ga-3) Loads ("gathers") elements of the type 'T' from Shared Local Memory locations addressed by byte offsets byte_offsets, and returns the loaded elements. More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::slm_gather (simd< uint32_t, N > byte_offsets, simd_mask< N > mask, simd< T, N > pass_thru, PropertyListT props={}) |
template <typename T, int N, typename PropertyListT = empty_properties_t> simd<T, N> slm_gather(simd<uint32_t, N> byte_offsets, simd_mask<N> mask, simd<T, N> pass_thru, PropertyListT props = {}); // (slm-ga-4) Loads ("gathers") elements of the type 'T' from Shared Local Memory locations addressed by byte offsets byte_offsets, and returns the loaded elements. More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::slm_gather (simd< uint32_t, N > byte_offsets, simd_mask< N > mask, PropertyListT props={}) |
template <typename T, int N, typename PropertyListT = empty_properties_t> simd<T, N> slm_gather(simd<uint32_t, N> byte_offsets, simd_mask<N> mask, PropertyListT props = {});// (slm-ga-5) Loads ("gathers") elements of the type 'T' from Shared Local Memory locations addressed by byte offsets byte_offsets, and returns the loaded elements. More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::slm_gather (simd< uint32_t, N > byte_offsets, PropertyListT props={}) |
template <typename T, int N, typename PropertyListT = empty_properties_t> simd<T, N> slm_gather(simd<uint32_t, N> byte_offsets, PropertyListT props = {}); // (slm-ga-6) Loads ("gathers") elements of the type 'T' from Shared Local Memory locations addressed by byte offsets byte_offsets, and returns the loaded elements. More... | |
| template<typename T , int N, int VS = 1, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::slm_gather (OffsetSimdViewT byte_offsets, simd_mask< N/VS > mask, simd< T, N > pass_thru, PropertyListT props={}) |
template <typename T, int N, int VS = 1, typename OffsetSimdViewT, typename PropertyListT = empty_props_t> simd <T, N> slm_gather( OffsetSimdViewT byte_offsets, simd_mask<N / VS> mask, simd<T, N> pass_thru, PropertyListT props = {}); // (slm-ga-7) Loads ("gathers") elements of the type 'T' from Shared Local Memory locations addressed by byte offsets byte_offsets, and returns the loaded elements. More... | |
| template<typename T , int N, int VS = 1, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::slm_gather (OffsetSimdViewT byte_offsets, simd_mask< N/VS > mask, PropertyListT props={}) |
simd <T, N> slm_gather( OffsetSimdViewT byte_offsets, simd_mask<N / VS> mask, PropertyListT props = {}); // (slm-ga-8) Loads ("gathers") elements of the type 'T' from Shared Local Memory locations addressed by byte offsets byte_offsets, and returns the loaded elements. More... | |
| template<typename T , int N, int VS = 1, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::slm_gather (OffsetSimdViewT byte_offsets, PropertyListT props={}) |
simd <T, N> slm_gather( OffsetSimdViewT byte_offsets, PropertyListT props = {}); // (slm-ga-9) Loads ("gathers") elements of the type 'T' from Shared Local Memory locations addressed by byte offsets byte_offsets, and returns the loaded elements. More... | |
| template<typename T > | |
| __ESIMD_API T | sycl::_V1::ext::intel::esimd::slm_scalar_load (uint32_t offset) |
| Load a scalar value from the Shared Local Memory. More... | |
| template<typename T , int N, int VS = 1, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::slm_scatter (simd< uint32_t, N/VS > byte_offsets, simd< T, N > vals, simd_mask< N/VS > mask, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename PropertyListT = empty_properties_t> void slm_scatter(simd<uint32_t, N / VS> byte_offsets, simd<T, N> vals, simd_mask<N / VS> mask, PropertyListT props = {}); // (slm-sc-1) void slm_scatter(simd<uint32_t, N / VS> byte_offsets, simd<T, N> vals, PropertyListT props = {}); // (slm-sc-2) More... | |
| template<typename T , int N, int VS = 1, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::slm_scatter (simd< uint32_t, N/VS > byte_offsets, simd< T, N > vals, PropertyListT props={}) |
template <typename T, int N, int VS = 1, typename PropertyListT = empty_properties_t> void slm_scatter(simd<uint32_t, N / VS> byte_offsets, simd<T, N> vals, PropertyListT props = {}); // (slm-sc-2) Stores ("scatters") elements of the type 'T' to Shared Local Memory locations addressed by byte offsets byte_offsets. More... | |
| template<typename T , int N, int VS = 1, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::slm_scatter (OffsetSimdViewT byte_offsets, simd< T, N > vals, simd_mask< N/VS > mask, PropertyListT props={}) |
template <typename T, int N, int VS = 1, typename OffsetSimdViewT, typename PropertyListT = empty_props_t> void slm_scatter( OffsetSimdViewT byte_offsets, simd<T, N> vals, simd_mask<N / VS> mask, PropertyListT props = {}); // (slm-sc-3) Stores ("scatters") elements of the type 'T' to Shared Local Memory locations addressed by byte offsets byte_offsets. More... | |
| template<typename T , int N, int VS = 1, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::slm_scatter (OffsetSimdViewT byte_offsets, simd< T, N > vals, PropertyListT props={}) |
void slm_scatter( OffsetSimdViewT byte_offsets, simd<T, N> vals, PropertyListT props = {}); // (slm-sc-4) Stores ("scatters") elements of the type 'T' to Shared Local Memory locations addressed by byte offsets byte_offsets. More... | |
| template<typename T > | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::slm_scalar_store (uint32_t offset, T val) |
| Store a scalar value into the Shared Local Memory. More... | |
| template<typename T , int N, rgba_channel_mask RGBAMask> | |
| __ESIMD_API std::enable_if_t<(N==8||N==16||N==32) &&(sizeof(T)==4), simd< T, N *get_num_channels_enabled(RGBAMask)> > | sycl::_V1::ext::intel::esimd::slm_gather_rgba (simd< uint32_t, N > offsets, simd_mask< N > mask=1) |
Gather data from the Shared Local Memory at specified offsets and return it as simd vector. More... | |
| template<typename T , int N, rgba_channel_mask Mask> | |
| __ESIMD_API std::enable_if_t<(N==8||N==16||N==32) &&(sizeof(T)==4)> | sycl::_V1::ext::intel::esimd::slm_scatter_rgba (simd< uint32_t, N > offsets, simd< T, N *get_num_channels_enabled(Mask)> vals, simd_mask< N > mask=1) |
Gather data from the Shared Local Memory at specified offsets and return it as simd vector. More... | |
| template<typename T , int N, typename Flags = overaligned_tag<detail::OperandSize::OWORD>> | |
| __ESIMD_API std::enable_if_t< is_simd_flag_type_v< Flags >, simd< T, N > > | sycl::_V1::ext::intel::esimd::slm_block_load (uint32_t byte_offset, Flags) |
Loads a contiguous block of SLM memory referenced by the given byte-offset offset, then returns the loaded data as a simd object. More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::slm_block_load (uint32_t byte_offset, PropertyListT props={}) |
Each of the following slm_block_load functions loads a contiguous memory block from SLM (Shared Local Memory) and the byte_offset. More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::slm_block_load (uint32_t byte_offset, simd_mask< 1 > pred, PropertyListT props={}) |
simd<T, N> slm_block_load(uint32_t byte_offset, simd_mask<N> pred, props = {}); // (slm-bl-2) Loads a contiguous memory block from SLM (Shared Local Memory) at the given byte_offset. More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::slm_block_load (uint32_t offset, simd_mask< 1 > pred, simd< T, N > pass_thru, PropertyListT props={}) |
simd<T, N> slm_block_load(uint32_t byte_offset, simd_mask<1> pred, simd<T, N> pass_thru, props={}); // (slm-bl-3) Loads a contiguous memory block from SLM (Shared Local Memory) at the given byte_offset. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (AccessorT lacc, uint32_t byte_offset, PropertyListT props={}) |
simd<T, N> block_load(local_accessor lacc, uint32_t byte_offset, props={}); // (lacc-bl-1) Loads a contiguous memory block from SLM (Shared Local Memory) associated with the local accessor lacc at the given byte_offset. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (AccessorT lacc, PropertyListT props={}) |
simd<T, N> block_load(local_accessor lacc, props={}); // (lacc-bl-2) Loads a contiguous memory block from SLM (Shared Local Memory) associated with the local accessor lacc at zero offset. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (AccessorT lacc, uint32_t byte_offset, simd_mask< 1 > pred, PropertyListT props={}) |
simd<T, N> block_load(local_accessor lacc, uint32_t byte_offset, simd_mask<1> pred, props={}); // (lacc-bl-3) Loads a contiguous memory block from SLM (Shared Local Memory) associated the local accessor lacc at the given byte_offset. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (AccessorT lacc, simd_mask< 1 > pred, PropertyListT props={}) |
simd<T, N> block_load(local_accessor lacc, simd_mask<1> pred, props={}); // (lacc-bl-4) Loads a contiguous memory block from SLM (Shared Local Memory) associated with the local accessor lacc at zero offset. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (AccessorT lacc, uint32_t byte_offset, simd_mask< 1 > pred, simd< T, N > pass_thru, PropertyListT props={}) |
simd<T, N> block_load(local_accessor lacc, uint32_t byte_offset, simd_mask<1> pred, simd<T, N> pass_thru, props={}); // (lacc-bl-5) Loads a contiguous memory block from SLM (Shared Local Memory) associated the local accessor lacc at the given byte_offset. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (AccessorT lacc, simd_mask< 1 > pred, simd< T, N > pass_thru, PropertyListT props={}) |
simd<T, N> block_load(local_accessor lacc, simd_mask<1> pred, simd<T, N> pass_thru, props={}); // (lacc-bl-6) Loads a contiguous memory block from SLM (Shared Local Memory) associated with the local accessor lacc at zero offset. More... | |
| template<typename T , int N, typename Flags > | |
| __ESIMD_API std::enable_if_t< is_simd_flag_type_v< Flags > > | sycl::_V1::ext::intel::esimd::slm_block_store (uint32_t offset, simd< T, N > vals, Flags) |
Stores elements of the vector vals to a contiguous block of SLM memory at the given byte-offset offset. More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::slm_block_store (uint32_t byte_offset, simd< T, N > vals, simd_mask< 1 > pred, PropertyListT props={}) |
Each of the following slm_block_store functions stores the vector vals to a contiguous memory block in SLM (Shared Local Memory) at the byte_offset. More... | |
| template<typename T , int N, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::slm_block_store (uint32_t byte_offset, simd< T, N > vals, PropertyListT props={}) |
void slm_block_store(uint32_t byte_offset, simd<T, N> vals, // (slm-bs-2) props = {}); Stores the vector vals to a contiguous memory block in SLM (Shared Local Memory) at the given byte_offset. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_write > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::block_store (AccessorT lacc, uint32_t byte_offset, simd< T, N > vals, PropertyListT props={}) |
void block_store(local_accessor lacc, uint32_t byte_offset, // (lacc-bs-1) simd<T, N> vals, props={}); Stores the vector vals to a contiguous memory block in SLM (Shared Local Memory) associated with the local accessor lacc at the given byte_offset. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_write > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::block_store (AccessorT lacc, simd< T, N > vals, PropertyListT props={}) |
void block_store(local_accessor lacc, simd<T, N> vals, // (lacc-bs-2) props={}); Stores the vector vals to a contiguous memory block in SLM (Shared Local Memory) associated with the local accessor lacc. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_write > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::block_store (AccessorT lacc, uint32_t byte_offset, simd< T, N > vals, simd_mask< 1 > pred, PropertyListT props={}) |
| void block_store(local_accessor lacc, uint32_t byte_offset, // (lacc-bs-3) simd<T, N> vals, simd_mask<1> pred, props={}); More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_write > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::block_store (AccessorT lacc, simd< T, N > vals, simd_mask< 1 > pred, PropertyListT props={}) |
void block_store(local_accessor lacc, simd<T, N> vals, // (lacc-bs-4) simd_mask<1> pred, props={}); Stores the vector vals to a contiguous memory block in SLM (Shared Local Memory) associated with the local accessor lacc. More... | |
| template<typename T , sycl::ext::intel::esimd::atomic_op Op> | |
| constexpr int | sycl::_V1::ext::intel::esimd::detail::lsc_to_internal_atomic_op () |
| template<atomic_op Op, typename T , int N, lsc_data_size DS> | |
| __ESIMD_API std::enable_if_t< get_num_args< Op >)==0, simd< T, N > > | sycl::_V1::ext::intel::esimd::detail::slm_atomic_update_impl (simd< uint32_t, N > offsets, simd_mask< N > pred) |
| SLM atomic. More... | |
| template<atomic_op Op, typename T , int N, lsc_data_size DS> | |
| __ESIMD_API std::enable_if_t< get_num_args< Op >)==1, simd< T, N > > | sycl::_V1::ext::intel::esimd::detail::slm_atomic_update_impl (simd< uint32_t, N > offsets, simd< T, N > src0, simd_mask< N > pred) |
| SLM atomic. More... | |
| template<atomic_op Op, typename T , int N, lsc_data_size DS> | |
| __ESIMD_API simd< T, N > | sycl::_V1::ext::intel::esimd::detail::slm_atomic_update_impl (simd< uint32_t, N > offsets, simd< T, N > src0, simd< T, N > src1, simd_mask< N > pred) |
| SLM atomic. More... | |
| template<atomic_op Op, typename T , int N> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==0, simd< T, N > > | sycl::_V1::ext::intel::esimd::slm_atomic_update (simd< uint32_t, N > byte_offset, simd_mask< N > mask=1) |
| More... | |
| template<atomic_op Op, typename T , int N, typename AccessorT > | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==0 &&__ESIMD_DNS::is_rw_local_accessor_v< AccessorT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorT lacc, simd< uint32_t, N > byte_offset, simd_mask< N > mask=1) |
simd<T, N> atomic_update(local_accessor lacc, simd<uint32_t, N> byte_offset, simd_mask<N> pred = 1); // (lacc-au0-1) Atomically updates N memory locations in SLM ssociated with the local accessor lacc at the given byte_offset, and returns a vector of old values found at the memory locations before update. More... | |
| template<atomic_op Op, typename T , int N> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==1, simd< T, N > > | sycl::_V1::ext::intel::esimd::slm_atomic_update (simd< uint32_t, N > byte_offset, simd< T, N > src0, simd_mask< N > mask=1) |
| One argument variant of the atomic update operation. More... | |
| template<atomic_op Op, typename T , int N, typename AccessorT > | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==1 &&__ESIMD_DNS::is_rw_local_accessor_v< AccessorT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorT lacc, simd< uint32_t, N > byte_offset, simd< T, N > src0, simd_mask< N > mask=1) |
| simd<T, N> atomic_update(local_accessor lacc, simd<uint32_t, N> byte_offset, simd<T, N> src0, simd_mask<1> pred = 1); // (lacc-au1-1) More... | |
| template<atomic_op Op, typename T , int N> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==2, simd< T, N > > | sycl::_V1::ext::intel::esimd::slm_atomic_update (simd< uint32_t, N > byte_offset, simd< T, N > src0, simd< T, N > src1, simd_mask< N > mask=1) |
| Two argument variant of the atomic update operation. More... | |
| template<atomic_op Op, typename T , int N, typename AccessorT > | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==2 &&__ESIMD_DNS::is_rw_local_accessor_v< AccessorT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorT lacc, simd< uint32_t, N > byte_offset, simd< T, N > src0, simd< T, N > src1, simd_mask< N > mask=1) |
| simd<T, N> atomic_update(local_accessor lacc, simd<uint32_t, N> byte_offset, simd<T, N> src0, simd<T, N> src1, simd_mask<1> pred = 1); // (lacc-au2-1) More... | |
| template<atomic_op Op, typename T , int N, lsc_data_size DS, typename PropertyListT , typename Toffset > | |
| __ESIMD_API std::enable_if_t< get_num_args< Op >)==0, simd< T, N > > | sycl::_V1::ext::intel::esimd::detail::atomic_update_impl (T *p, simd< Toffset, N > offsets, simd_mask< N > pred) |
| USM pointer atomic. More... | |
| template<atomic_op Op, typename T , int N, lsc_data_size DS, typename PropertyListT , typename Toffset > | |
| __ESIMD_API std::enable_if_t< get_num_args< Op >)==1, simd< T, N > > | sycl::_V1::ext::intel::esimd::detail::atomic_update_impl (T *p, simd< Toffset, N > offsets, simd< T, N > src0, simd_mask< N > pred) |
| USM pointer atomic. More... | |
| template<atomic_op Op, typename T , int N, lsc_data_size DS, typename PropertyListT , typename Toffset > | |
| __ESIMD_API std::enable_if_t< get_num_args< Op >)==2, simd< T, N > > | sycl::_V1::ext::intel::esimd::detail::atomic_update_impl (T *p, simd< Toffset, N > offsets, simd< T, N > src0, simd< T, N > src1, simd_mask< N > pred) |
| USM pointer atomic. More... | |
| template<atomic_op Op, typename T , int N, lsc_data_size DS = lsc_data_size::default_size, typename PropertyListT , typename AccessorTy , typename Toffset > | |
| __ESIMD_API std::enable_if_t< get_num_args< Op >)==0 &&__ESIMD_DNS::is_rw_device_accessor_v< AccessorTy >, simd< T, N > > | sycl::_V1::ext::intel::esimd::detail::atomic_update_impl (AccessorTy acc, simd< Toffset, N > byte_offsets, simd_mask< N > pred) |
| Accessor-based atomic. More... | |
| template<atomic_op Op, typename T , int N, lsc_data_size DS, typename PropertyListT , typename AccessorTy , typename Toffset > | |
| __ESIMD_API std::enable_if_t< get_num_args< Op >)==1 &&__ESIMD_DNS::is_rw_device_accessor_v< AccessorTy >, simd< T, N > > | sycl::_V1::ext::intel::esimd::detail::atomic_update_impl (AccessorTy acc, simd< Toffset, N > byte_offset, simd< T, N > src0, simd_mask< N > pred) |
| Accessor-based atomic. More... | |
| template<atomic_op Op, typename T , int N, lsc_data_size DS, typename PropertyListT , typename AccessorTy , typename Toffset > | |
| __ESIMD_API std::enable_if_t< get_num_args< Op >)==2 &&__ESIMD_DNS::is_rw_device_accessor_v< AccessorTy >, simd< T, N > > | sycl::_V1::ext::intel::esimd::detail::atomic_update_impl (AccessorTy acc, simd< Toffset, N > byte_offset, simd< T, N > src0, simd< T, N > src1, simd_mask< N > pred) |
| Accessor-based atomic. More... | |
| template<atomic_op Op, typename T , int N, typename Toffset , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==0 &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (T *p, simd< Toffset, N > byte_offset, simd_mask< N > mask, PropertyListT props={}) |
| More... | |
| template<atomic_op Op, typename T , int N, typename Toffset , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==0 &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (T *p, simd< Toffset, N > byte_offset, PropertyListT props={}) |
| simd<T, N> atomic_update(T *p, simd<Toffset, N> byte_offset, props = {}); /// (usm-au0-2) More... | |
| template<atomic_op Op, typename T , int N, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==0 &&ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_simd_view_type_v< OffsetSimdViewT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (T *p, OffsetSimdViewT offsets, simd_mask< N > mask, PropertyListT props={}) |
| simd<T, N> atomic_update(T *p, OffsetSimdViewT byte_offset, simd_mask<N> mask, props = {}); /// (usm-au0-3) More... | |
| template<atomic_op Op, typename T , int N, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==0 &&ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_simd_view_type_v< OffsetSimdViewT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (T *p, OffsetSimdViewT byte_offset, PropertyListT props={}) |
| simd<T, N> atomic_update(T *p, OffsetSimdViewT byte_offset, props = {}); /// (usm-au0-4) More... | |
| template<atomic_op Op, typename T , int N, typename Toffset > | |
| __ESIMD_API std::enable_if_t< std::is_integral_v< Toffset >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (T *p, Toffset byte_offset, simd_mask< N > mask=1) |
A variation of atomic_update API with offset represented as scalar. More... | |
| template<atomic_op Op, typename T , int N, typename Toffset , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==1 &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (T *p, simd< Toffset, N > byte_offset, simd< T, N > src0, simd_mask< N > mask, PropertyListT props={}) |
| More... | |
| template<atomic_op Op, typename T , int N, typename Toffset , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==1 &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (T *p, simd< Toffset, N > byte_offset, simd< T, N > src0, PropertyListT props={}) |
| simd<T, N> atomic_update(T *ptr, simd<Toffset, N> byte_offset, simd<T, N> src0, props = {}); // (usm-au1-2) More... | |
| template<atomic_op Op, typename T , int N, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==1 &&ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_simd_view_type_v< OffsetSimdViewT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (T *p, OffsetSimdViewT offsets, simd< T, N > src0, simd_mask< N > mask, PropertyListT props={}) |
| simd<T, N> atomic_update(T *p, OffsetSimdViewT byte_offset, simd<T, N> src0, simd_mask<N> mask, props = {}); // (usm-au1-3) More... | |
| template<atomic_op Op, typename T , int N, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==1 &&ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_simd_view_type_v< OffsetSimdViewT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (T *p, OffsetSimdViewT offsets, simd< T, N > src0, PropertyListT props={}) |
| simd<T, N> atomic_update(T *p, OffsetSimdViewT byte_offset, simd<T, N> src0, props = {}); // (usm-au1-4) More... | |
| template<atomic_op Op, typename Tx , int N, typename Toffset > | |
| __ESIMD_API std::enable_if_t< std::is_integral_v< Toffset > &&((Op !=atomic_op::store &&Op !=atomic_op::xchg)||N==1), simd< Tx, N > > | sycl::_V1::ext::intel::esimd::atomic_update (Tx *p, Toffset byte_offset, simd< Tx, N > src0, simd_mask< N > mask) |
A variation of atomic_update API with offset represented as scalar object. More... | |
| template<atomic_op Op, typename T , int N, typename Toffset , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==2 &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (T *p, simd< Toffset, N > byte_offset, simd< T, N > src0, simd< T, N > src1, simd_mask< N > mask, PropertyListT props={}) |
Atomically updates N memory locations represented by a USM pointer and a vector of offsets relative to the pointer, and returns a vector of old values found at the memory locations before update. More... | |
| template<atomic_op Op, typename T , int N, typename Toffset , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==2 &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (T *p, simd< Toffset, N > byte_offset, simd< T, N > src0, simd< T, N > src1, PropertyListT props={}) |
| simd<T, N> atomic_update(T *p, simd<Toffset, N> byte_offset, simd<T, N> src0, simd<T, N> src1, props = {}); // (usm-au2-2) More... | |
| template<atomic_op Op, typename T , int N, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==2 &&ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_simd_view_type_v< OffsetSimdViewT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (T *p, OffsetSimdViewT byte_offset, simd< T, N > src0, simd< T, N > src1, simd_mask< N > mask, PropertyListT props={}) |
| simd<T, N> atomic_update(T *p, OffsetSimdViewT byte_offset, simd<T, N> src0, simd<T, N> src1, simd_mask<N> mask, props = {}) // (usm-au2-3) More... | |
| template<atomic_op Op, typename T , int N, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==2 &&ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_simd_view_type_v< OffsetSimdViewT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (T *p, OffsetSimdViewT byte_offset, simd< T, N > src0, simd< T, N > src1, PropertyListT props={}) |
| simd<T, N> atomic_update(T *p, OffsetSimdViewT byte_offset, simd<T, N> src0, simd<T, N> src1, props = {}) // (usm-au2-4) More... | |
| template<atomic_op Op, typename Tx , int N, typename Toffset > | |
| __ESIMD_API std::enable_if_t< std::is_integral_v< Toffset >, simd< Tx, N > > | sycl::_V1::ext::intel::esimd::atomic_update (Tx *p, Toffset byte_offset, simd< Tx, N > src0, simd< Tx, N > src1, simd_mask< N > mask) |
A variation of atomic_update API with byte_offset represented as scalar. More... | |
| template<atomic_op Op, typename T , int N, typename Toffset , typename AccessorTy , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==0 &&__ESIMD_DNS::is_rw_device_accessor_v< AccessorTy > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, simd< Toffset, N > byte_offset, simd_mask< N > mask, PropertyListT props={}) |
| More... | |
| template<atomic_op Op, typename T , int N, typename Toffset , typename AccessorTy , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==0 &&__ESIMD_DNS::is_rw_device_accessor_v< AccessorTy > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, simd< Toffset, N > byte_offset, PropertyListT props={}) |
simd<T, N> atomic_update(AccessorT acc, simd<Toffset, N> byte_offset, props = {}); /// (acc-au0-2) A variation of atomic_update API without mask operand More... | |
| template<atomic_op Op, typename T , int N, typename OffsetSimdViewT , typename AccessorTy , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==0 &&__ESIMD_DNS::is_rw_device_accessor_v< AccessorTy > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_simd_view_type_v< OffsetSimdViewT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, OffsetSimdViewT byte_offset, simd_mask< N > mask, PropertyListT props={}) |
simd<T, N> atomic_update(AccessorT acc, OffsetSimdViewT byte_offset, simd_mask<N> mask, props = {}); /// (acc-au0-3) A variation of atomic_update API with offsets represented as simd_view object. More... | |
| template<atomic_op Op, typename T , int N, typename OffsetSimdViewT , typename AccessorTy , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==0 &&__ESIMD_DNS::is_rw_device_accessor_v< AccessorTy > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_simd_view_type_v< OffsetSimdViewT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, OffsetSimdViewT byte_offset, PropertyListT props={}) |
simd<T, N> atomic_update(AccessorT acc, OffsetSimdViewT byte_offset, props = {}); /// (acc-au0-4) A variation of atomic_update API with offsets represented as simd_view object and no mask operand. More... | |
| template<atomic_op Op, typename T , int N, typename Toffset , typename AccessorTy > | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==0 &&__ESIMD_DNS::is_rw_device_accessor_v< AccessorTy >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, Toffset byte_offset, simd_mask< N > mask) |
A variation of atomic_update API with offset represented as scalar. More... | |
| template<atomic_op Op, typename T , int N, typename AccessorTy > | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==0 &&__ESIMD_DNS::is_rw_local_accessor_v< AccessorTy >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, uint32_t byte_offset, simd_mask< N > mask) |
A variation of atomic_update API with byte_offset represented as scalar using local_accessor. More... | |
| template<atomic_op Op, typename T , int N, typename Toffset , typename AccessorTy , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==1 &&__ESIMD_DNS::is_rw_device_accessor_v< AccessorTy > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, simd< Toffset, N > byte_offset, simd< T, N > src0, simd_mask< N > mask, PropertyListT props={}) |
| More... | |
| template<atomic_op Op, typename T , int N, typename Toffset , typename AccessorTy , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==1 &&__ESIMD_DNS::is_rw_device_accessor_v< AccessorTy > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, simd< Toffset, N > byte_offset, simd< T, N > src0, PropertyListT props={}) |
| simd<T, N> atomic_update(AccessorT acc, simd<Toffset, N> byte_offset, simd<T, N> src0, props = {}); // (acc-au1-2) More... | |
| template<atomic_op Op, typename T , int N, typename OffsetSimdViewT , typename AccessorTy , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==1 &&__ESIMD_DNS::is_rw_device_accessor_v< AccessorTy > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_simd_view_type_v< OffsetSimdViewT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, OffsetSimdViewT byte_offset, simd< T, N > src0, simd_mask< N > mask, PropertyListT props={}) |
| simd<T, N> atomic_update(AccessorT acc, OffsetSimdViewT byte_offset, simd<T, N> src0, simd_mask<N> mask, props = {}); // (acc-au1-3) More... | |
| template<atomic_op Op, typename T , int N, typename OffsetSimdViewT , typename AccessorTy , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==1 &&__ESIMD_DNS::is_rw_device_accessor_v< AccessorTy > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_simd_view_type_v< OffsetSimdViewT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, OffsetSimdViewT byte_offset, simd< T, N > src0, PropertyListT props={}) |
| simd<T, N> atomic_update(AccessorT acc, OffsetSimdViewT byte_offset, simd<T, N> src0, props = {}); // (acc-au1-4) More... | |
| template<atomic_op Op, typename T , int N, typename Toffset , typename AccessorTy > | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::is_rw_device_accessor_v< AccessorTy > &&((Op !=atomic_op::store &&Op !=atomic_op::xchg)||N==1), simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, Toffset offset, simd< T, N > src0, simd_mask< N > mask) |
A variation of atomic_update API with offset represented as scalar object. More... | |
| template<atomic_op Op, typename Tx , int N, typename AccessorTy > | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::is_rw_local_accessor_v< AccessorTy > &&((Op !=atomic_op::store &&Op !=atomic_op::xchg)||N==1), simd< Tx, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, uint32_t offset, simd< Tx, N > src0, simd_mask< N > mask) |
A variation of atomic_update API with offset represented as scalar object and uses local_accessor. More... | |
| template<atomic_op Op, typename T , int N, typename Toffset , typename AccessorTy , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==2 &&std::is_integral_v< Toffset > &&__ESIMD_DNS::is_rw_device_accessor_v< AccessorTy > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, simd< Toffset, N > byte_offset, simd< T, N > src0, simd< T, N > src1, simd_mask< N > mask, PropertyListT props={}) |
| More... | |
| template<atomic_op Op, typename T , int N, typename Toffset , typename AccessorTy , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==2 &&__ESIMD_DNS::is_rw_device_accessor_v< AccessorTy > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, simd< Toffset, N > byte_offset, simd< T, N > src0, simd< T, N > src1, PropertyListT props={}) |
| simd<T, N> atomic_update(AccessorTy acc, simd<Toffset, N> byte_offset, simd<T, N> src0, simd<T, N> src1, props = {}); // (acc-au2-2) More... | |
| template<atomic_op Op, typename T , int N, typename OffsetSimdViewT , typename AccessorTy , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==2 &&__ESIMD_DNS::is_rw_device_accessor_v< AccessorTy > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_simd_view_type_v< OffsetSimdViewT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, OffsetSimdViewT byte_offset, simd< T, N > src0, simd< T, N > src1, simd_mask< N > mask, PropertyListT props={}) |
| simd<T, N> atomic_update(AccessorTy acc, OffsetSimdViewT byte_offset, simd<T, N> src0, simd<T, N> src1, simd_mask<N> mask, props = {}); // (acc-au2-3) More... | |
| template<atomic_op Op, typename T , int N, typename OffsetSimdViewT , typename AccessorTy , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==2 &&__ESIMD_DNS::is_rw_device_accessor_v< AccessorTy > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > &&detail::is_simd_view_type_v< OffsetSimdViewT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, OffsetSimdViewT byte_offset, simd< T, N > src0, simd< T, N > src1, PropertyListT props={}) |
| simd<T, N> atomic_update(AccessorTy acc, OffsetSimdViewT, byte_offset, simd<T, N> src0, simd<T, N> src1, props = {}); // (acc-au2-4) More... | |
| template<atomic_op Op, typename Tx , int N, typename Toffset , typename AccessorTy > | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::is_rw_device_accessor_v< AccessorTy >, simd< Tx, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, Toffset offset, simd< Tx, N > src0, simd< Tx, N > src1, simd_mask< N > mask) |
A variation of atomic_update API with offsets represented as scalar. More... | |
| template<atomic_op Op, typename Tx , int N, typename AccessorTy > | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::is_rw_local_accessor_v< AccessorTy >, simd< Tx, N > > | sycl::_V1::ext::intel::esimd::atomic_update (AccessorTy acc, uint32_t offset, simd< Tx, N > src0, simd< Tx, N > src1, simd_mask< N > mask) |
A variation of atomic_update API with offsets represented as scalar and local_accessor is used. More... | |
| template<uint8_t cntl> | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::fence () |
| esimd::fence sets the memory read/write order. More... | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::fence (fence_mask cntl) |
| template<memory_kind Kind = memory_kind::global, fence_flush_op FenceOp = fence_flush_op::none, fence_scope Scope = fence_scope::group> | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::fence () |
| Memory fence. More... | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::barrier () |
| Generic work-group barrier. More... | |
| template<typename T , int m, int N, typename AccessorTy , unsigned plane = 0> | |
| __ESIMD_API simd< T, m *N > | sycl::_V1::ext::intel::esimd::media_block_load (AccessorTy acc, unsigned x, unsigned y) |
| Media block load. More... | |
| template<typename T , int m, int N, typename AccessorTy , unsigned plane = 0> | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::media_block_store (AccessorTy acc, unsigned x, unsigned y, simd< T, m *N > vals) |
| Media block store. More... | |
| template<typename T , int N, typename AccessorTy , typename Flags = overaligned_tag<detail::OperandSize::OWORD>> | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorTy, detail::accessor_mode_cap::can_read > &&is_simd_flag_type_v< Flags >, simd< T, N > > | sycl::_V1::ext::intel::esimd::block_load (AccessorTy acc, uint32_t byte_offset, Flags flags) |
Loads a contiguous block of SLM memory referenced by the given local-accessor acc and byte_offset, then returns the loaded data as a simd object. More... | |
| template<typename T , int N, typename AccessorT , typename Flags > | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_write > &&is_simd_flag_type_v< Flags > > | sycl::_V1::ext::intel::esimd::block_store (AccessorT acc, uint32_t offset, simd< T, N > vals, Flags flags) |
| Variant of block_store that uses local accessor as a parameter. More... | |
| template<typename T , int N, int VS, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, simd< uint32_t, N/VS > byte_offsets, simd_mask< N/VS > mask, simd< T, N > pass_thru, PropertyListT props={}) |
| Variant of gather that uses local accessor as a parameter template <typename T, int N, int VS, typename AccessorT, typename PropertyListT = empty_properties_t> simd<T, N> gather(AccessorT acc, simd<uint32_t, N / VS> byte_offsets, simd_mask<N / VS> mask, simd<T, N> pass_thru, PropertyListT props = {}); // (lacc-ga-1) simd<T, N> gather(AccessorT acc, simd<uint32_t, N / VS> byte_offsets, simd_mask<N / VS> mask, PropertyListT props = {}); // (lacc-ga-2) simd<T, N> gather(AccessorT acc, simd<uint32_t, N / VS> byte_offsets, PropertyListT props = {}); // (lacc-ga-3) More... | |
| template<typename T , int N, int VS, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, simd< uint32_t, N/VS > byte_offsets, simd_mask< N/VS > mask, PropertyListT props={}) |
| template <typename T, int N, int VS, typename AccessorT, typename PropertyListT = empty_properties_t> simd<T, N> gather(AccessorT acc, simd<uint32_t, N / VS> byte_offsets, simd_mask<N / VS> mask, PropertyListT props = {}); // (lacc-ga-2) Supported platforms: DG2, PVC in most cases. More... | |
| template<typename T , int N, int VS, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, simd< uint32_t, N/VS > byte_offsets, PropertyListT props={}) |
| template <typename T, int N, int VS, typename AccessorT, typename PropertyListT = empty_properties_t> simd<T, N> gather(AccessorT acc, simd<uint32_t, N / VS> byte_offsets, PropertyListT props = {}); // (lacc-ga-3) Supported platforms: DG2, PVC in most cases. More... | |
| template<typename T , int N, typename AccessorT , typename MaskT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&std::is_same_v< MaskT, simd_mask< N >> &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, simd< uint32_t, N > byte_offsets, MaskT mask, simd< T, N > pass_thru, PropertyListT props={}) |
| template <typename T, int N, typename AccessorT, typename PropertyListT = empty_properties_t> simd<T, N> gather(AccessorT acc, simd<uint32_t, N> byte_offsets, simd_mask<N> mask, simd<T, N> pass_thru, PropertyListT props = {}); // (lacc-ga-4) This function is identical to (lacc-ga-1) except that vector size is fixed to 1. More... | |
| template<typename T , int N, typename AccessorT , typename MaskT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&std::is_same_v< MaskT, simd_mask< N >> &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, simd< uint32_t, N > byte_offsets, MaskT mask, PropertyListT props={}) |
| template <typename T, int N, typename AccessorT, typename PropertyListT = empty_properties_t> simd<T, N> gather(AccessorT acc, simd<uint32_t, N> byte_offsets, simd_mask<N> mask, PropertyListT props // (lacc-ga-5) This function is identical to (lacc-ga-2) except that vector size is fixed to 1. More... | |
| template<typename T , int N, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, simd< uint32_t, N > byte_offsets, PropertyListT props={}) |
| template <typename T, int N, typename AccessorT, typename PropertyListT = empty_properties_t> simd<T, N> gather(AccessorT acc, simd<uint32_t, N> byte_offsets, PropertyListT props = {}); // (lacc-ga-6) This function is identical to (lacc-ga-3) except that vector size is fixed to 1. More... | |
| template<typename T , int N, int VS = 1, typename AccessorT , typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, OffsetSimdViewT byte_offsets, simd_mask< N/VS > mask, simd< T, N > pass_thru, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename OffsetSimdViewT, More... | |
| template<typename T , int N, int VS = 1, typename AccessorT , typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, OffsetSimdViewT byte_offsets, simd_mask< N/VS > mask, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename AccessorT, typename OffsetSimdViewT, More... | |
| template<typename T , int N, int VS = 1, typename AccessorT , typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t<(detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT >), simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorT acc, OffsetSimdViewT byte_offsets, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename AccessorT, typename OffsetSimdViewT, More... | |
| template<typename T , int N, typename AccessorTy > | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorTy, detail::accessor_mode_cap::can_read >, simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorTy acc, simd< uint32_t, N > offsets, uint32_t glob_offset, simd_mask< N > mask=1) |
| Variant of gather that uses local accessor as a parameter. More... | |
| template<typename T , int N, int VS = 1, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_write > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::scatter (AccessorT acc, simd< uint32_t, N/VS > byte_offsets, simd< T, N > vals, simd_mask< N/VS > mask, PropertyListT props={}) |
| Variant of scatter that uses local accessor as a parameter template <typename T, int N, int VS = 1, typename AccessorT, typename PropertyListT = empty_properties_t> void scatter(AccessorT acc, simd<uint32_t, N / VS> byte_offsets, simd<T, N> vals, simd_mask<N / VS> mask, PropertyListT props = {}); // (lacc-sc-1) More... | |
| template<typename T , int N, int VS = 1, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_write > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::scatter (AccessorT acc, simd< uint32_t, N/VS > byte_offsets, simd< T, N > vals, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename AccessorT, typename PropertyListT = empty_properties_t> void scatter(AccessorT acc, simd<uint32_t, N / VS> byte_offsets, simd<T, N> vals, PropertyListT props = {}); // (lacc-sc-2) More... | |
| template<typename T , int N, int VS = 1, typename OffsetSimdViewT , typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_write > &&detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::scatter (AccessorT acc, OffsetSimdViewT byte_offsets, simd< T, N > vals, simd_mask< N/VS > mask, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename AccessorT, typename OffsetSimdViewT, typename PropertyListT = empty_properties_t> void scatter(AccessorT acc, OffsetSimdViewT byte_offsets, simd<T, N> vals, simd_mask<N / VS> mask, PropertyListT props = {}); // (lacc-sc-3) More... | |
| template<typename T , int N, int VS = 1, typename OffsetSimdViewT , typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_write > &&detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::scatter (AccessorT acc, OffsetSimdViewT byte_offsets, simd< T, N > vals, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename OffsetSimdViewT, typename AccessorT, typename PropertyListT = empty_properties_t> void scatter(AccessorT acc, OffsetSimdViewT byte_offsets, simd<T, N> vals, PropertyListT props = {}); // (lacc-sc-4) More... | |
| template<typename T , int N, typename AccessorTy > | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorTy, detail::accessor_mode_cap::can_write > > | sycl::_V1::ext::intel::esimd::scatter (AccessorTy acc, simd< uint32_t, N > offsets, simd< T, N > vals, uint32_t glob_offset, simd_mask< N > mask=1) |
| Variant of scatter that uses local accessor as a parameter. More... | |
| template<typename T , int N, int VS, typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (const T *p, simd< OffsetT, N/VS > byte_offsets, simd_mask< N/VS > mask, PropertyListT props={}) |
| template <typename T, int N, int VS, typename OffsetT, typename PropertyListT = empty_properties_t> void prefetch(const T *p, simd<OffsetT, N / VS> byte_offsets, simd_mask<N / VS> mask, PropertyListT props = {}); // (usm-pf-1) void prefetch(const T *p, simd<OffsetT, N / VS> byte_offsets, PropertyListT props = {}); // (usm-pf-2) More... | |
| template<typename T , int N, int VS, typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (const T *p, simd< OffsetT, N/VS > byte_offsets, PropertyListT props={}) |
| template <typename T, int VS, typename OffsetT, typename PropertyListT = empty_properties_t> void prefetch(const T *p, simd<OffsetT, N / VS> byte_offsets, PropertyListT props = {}); // (usm-pf-2) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int N, typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (const T *p, simd< OffsetT, N > byte_offsets, simd_mask< N > mask, PropertyListT props={}) |
| template <typename T, int N, typename OffsetT, typename PropertyListT = empty_properties_t> void prefetch(const T *p, simd<OffsetT, N> byte_offsets, simd_mask<N> mask, PropertyListT props = {}); // (usm-pf-3) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int N, typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (const T *p, simd< OffsetT, N > byte_offsets, PropertyListT props={}) |
| template <typename T, typename OffsetT, typename PropertyListT = empty_properties_t> void prefetch(const T *p, simd<OffsetT, N> byte_offsets, PropertyListT props = {}); // (usm-pf-4) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int N, int VS = 1, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (const T *p, OffsetSimdViewT byte_offsets, simd_mask< N/VS > mask, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename OffsetSimdViewT, typename PropertyListT = empty_properties_t> void prefetch(const T *p, OffsetSimdViewT byte_offsets, simd_mask<N / VS> mask, PropertyListT props = {}); // (usm-pf-5) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int N, int VS = 1, typename OffsetSimdViewT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (const T *p, OffsetSimdViewT byte_offsets, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename OffsetSimdViewT, typename PropertyListT = empty_properties_t> void prefetch(const T *p, OffsetSimdViewT byte_offsets, PropertyListT props = {}); // (usm-pf-6) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int VS = 1, typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< std::is_integral_v< OffsetT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (const T *p, OffsetT byte_offset, simd_mask< 1 > mask, PropertyListT props={}) |
| template <typename T, int VS = 1, typename OffsetT, typename PropertyListT = empty_properties_t> void prefetch(const T *p, OffsetT byte_offset, simd_mask<1> mask, PropertyListT props = {}); // (usm-pf-7) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int VS = 1, typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< std::is_integral_v< OffsetT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (const T *p, OffsetT byte_offset, PropertyListT props={}) |
| template <typename T, int VS = 1, typename OffsetT, typename PropertyListT = empty_properties_t> void prefetch(const T *p, OffsetT byte_offset, PropertyListT props = {}); // (usm-pf-8) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int VS = 1, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (const T *p, simd_mask< 1 > mask, PropertyListT props={}) |
| template <typename T, int VS = 1, typename PropertyListT = empty_properties_t> void prefetch(const T *p, simd_mask<1> mask, PropertyListT props = {}); //(usm-pf-9) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int VS = 1, typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (const T *p, PropertyListT props={}) |
| template <typename T, int VS = 1, typename PropertyListT = empty_properties_t> void prefetch(const T *p, PropertyListT props = {}); // (usm-pf-10) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int N, int VS, typename AccessorT , typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (AccessorT acc, simd< OffsetT, N/VS > byte_offsets, simd_mask< N/VS > mask, PropertyListT props={}) |
| template <typename T, int N, int VS, typename AccessorT, typename OffsetT, typename PropertyListT = empty_properties_t> void prefetch(AccessorT acc, simd<OffsetT, N / VS> byte_offsets, simd_mask<N / VS> mask, PropertyListT props = {}); // (acc-pf-1) void prefetch(AccessorT acc, simd<OffsetT, N / VS> byte_offsets, PropertyListT props = {}); // (acc-pf-2) More... | |
| template<typename T , int N, int VS, typename AccessorT , typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (AccessorT acc, simd< OffsetT, N/VS > byte_offsets, PropertyListT props={}) |
| template <typename T, int N, int VS, typename AccessorT, typename OffsetT, typename PropertyListT = empty_properties_t> void prefetch(AccessorT acc, simd<uint32_t, N / VS> byte_offsets, PropertyListT props = {}); // (acc-pf-2) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int N, typename AccessorT , typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (AccessorT acc, simd< OffsetT, N > byte_offsets, simd_mask< N > mask, PropertyListT props={}) |
| template <typename T, int N, typename AccessorT, typename OffsetT, typename PropertyListT = empty_properties_t> void prefetch(AccessorT acc, simd<uint32_t, N> byte_offsets, simd_mask<N> mask, PropertyListT props = {}); // (acc-pf-3) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int N, typename AccessorT , typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (AccessorT acc, simd< OffsetT, N > byte_offsets, PropertyListT props={}) |
| template <typename T, typename AccessorT, typename OffsetT, typename PropertyListT = empty_properties_t> void prefetch(AccessorT acc, simd<uint32_t, N> byte_offsets, PropertyListT props = {}); // (acc-pf-4) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int N, int VS = 1, typename OffsetSimdViewT , typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (AccessorT acc, OffsetSimdViewT byte_offsets, simd_mask< N/VS > mask, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename OffsetSimdViewT, typename AccessorT, typename PropertyListT = empty_properties_t> void prefetch(AccessorT acc, OffsetSimdViewT byte_offsets, simd_mask<N / VS> mask, PropertyListT props = {}); // (acc-pf-5) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int N, int VS = 1, typename OffsetSimdViewT , typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&detail::is_simd_view_type_v< OffsetSimdViewT > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (AccessorT acc, OffsetSimdViewT byte_offsets, PropertyListT props={}) |
| template <typename T, int N, int VS = 1, typename OffsetSimdViewT, typename AccessorT, typename PropertyListT = empty_properties_t> void prefetch(AccessorT acc, OffsetSimdViewT byte_offsets, PropertyListT props = {}); // (acc-pf-6) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int VS = 1, typename AccessorT , typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< std::is_integral_v< OffsetT > &&detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (AccessorT acc, OffsetT byte_offset, simd_mask< 1 > mask, PropertyListT props={}) |
| template <typename T, int VS = 1, typename AccessorT, typename OffsetT, typename PropertyListT = empty_properties_t> void prefetch(AccessorT acc, uint32_t byte_offset, simd_mask<1> mask, PropertyListT props = {}); // (acc-pf-7) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int VS = 1, typename AccessorT , typename OffsetT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< std::is_integral_v< OffsetT > &&detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (AccessorT acc, OffsetT byte_offset, PropertyListT props={}) |
| template <typename T, int VS = 1, typename AccessorT, typename OffsetT, typename PropertyListT = empty_properties_t> void prefetch(AccessorT acc, uint32_t byte_offset, PropertyListT props = {}); // (acc-pf-8) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int VS = 1, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (AccessorT acc, simd_mask< 1 > mask, PropertyListT props={}) |
| template <typename T, int VS = 1, typename AccessorT, typename PropertyListT = empty_properties_t> void prefetch(AccessorT acc, simd_mask<1> mask, PropertyListT props = {}); //(acc-pf-9) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int VS = 1, typename AccessorT , typename PropertyListT = ext::oneapi::experimental::detail::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< detail::is_device_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read > &&ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch (AccessorT acc, PropertyListT props={}) |
| template <typename T, int VS = 1, typename AccessorT, typename PropertyListT = empty_properties_t> void prefetch(AccessorT acc, PropertyListT props = {}); // (acc-pf-10) Supported platforms: DG2, PVC only. More... | |
| template<typename T , int BlockWidth, int BlockHeight = 1, int NBlocks = 1, bool Transposed = false, bool Transformed = false, int N = detail::get_lsc_block_2d_data_size< T, NBlocks, BlockHeight, BlockWidth, Transposed, Transformed>(), typename PropertyListT = oneapi::experimental::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT >, simd< T, N > > | sycl::_V1::ext::intel::esimd::load_2d (const T *Ptr, unsigned SurfaceWidth, unsigned SurfaceHeight, unsigned SurfacePitch, int X, int Y, PropertyListT props={}) |
| template <typename T, int BlockWidth, int BlockHeight = 1, int NBlocks = 1, bool Transposed = false, bool Transformed = false, int N = detail::get_lsc_block_2d_data_size< T, NBlocks, BlockHeight, BlockWidth, Transposed, Transformed>(), typename PropertyListT = empty_properties_t> simd<T, N> load_2d(const T *Ptr, unsigned SurfaceWidth, unsigned SurfaceHeight, unsigned SurfacePitch, int X, int Y, PropertyListT props = {}); 2D USM pointer block load. More... | |
| template<typename T , int BlockWidth, int BlockHeight = 1, int NBlocks = 1, int N = detail::get_lsc_block_2d_data_size< T, NBlocks, BlockHeight, BlockWidth, false , false >(), typename PropertyListT = oneapi::experimental::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::prefetch_2d (const T *Ptr, unsigned SurfaceWidth, unsigned SurfaceHeight, unsigned SurfacePitch, int X, int Y, PropertyListT props={}) |
| template <typename T, int BlockWidth, int BlockHeight = 1, int NBlocks = 1, int N = detail::get_lsc_block_2d_data_size< T, NBlocks, BlockHeight, BlockWidth, false, false>(), typename PropertyListT = empty_properties_t> void prefetch_2d(const T *Ptr, unsigned SurfaceWidth, unsigned SurfaceHeight, unsigned SurfacePitch, int X, int Y, PropertyListT props = {}); 2D USM pointer block prefetch. More... | |
| template<typename T , int BlockWidth, int BlockHeight = 1, int N = detail::get_lsc_block_2d_data_size< T, 1u, BlockHeight, BlockWidth, false , false >(), typename PropertyListT = oneapi::experimental::empty_properties_t> | |
| __ESIMD_API std::enable_if_t< ext::oneapi::experimental::is_property_list_v< PropertyListT > > | sycl::_V1::ext::intel::esimd::store_2d (T *Ptr, unsigned SurfaceWidth, unsigned SurfaceHeight, unsigned SurfacePitch, int X, int Y, simd< T, N > Vals, PropertyListT props={}) |
| 2D USM pointer block store. More... | |
| template<rgba_channel_mask RGBAMask = rgba_channel_mask::ABGR, typename AccessorT , int N, typename T = typename AccessorT::value_type> | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_read >, simd< T, N *get_num_channels_enabled(RGBAMask)> > | sycl::_V1::ext::intel::esimd::gather_rgba (AccessorT acc, simd< uint32_t, N > offsets, uint32_t global_offset=0, simd_mask< N > mask=1) |
| Variant of gather_rgba that uses local accessor as a parameter. More... | |
| template<rgba_channel_mask RGBAMask = rgba_channel_mask::ABGR, typename AccessorT , int N, typename T = typename AccessorT::value_type> | |
| __ESIMD_API std::enable_if_t< detail::is_local_accessor_with_v< AccessorT, detail::accessor_mode_cap::can_write > > | sycl::_V1::ext::intel::esimd::scatter_rgba (AccessorT acc, simd< uint32_t, N > offsets, simd< T, N *get_num_channels_enabled(RGBAMask)> vals, uint32_t global_offset=0, simd_mask< N > mask=1) |
Variant of scatter_rgba that uses local accessor as a parameter Gather data from the memory addressed by accessor acc, offset common for all loaded elements global_offset and per-element offsets offsets, and return it as simd vector. More... | |
| template<uint8_t exec_size, uint8_t sfid, uint8_t num_src0, uint8_t num_src1, uint8_t num_dst, raw_send_eot eot = raw_send_eot::not_eot, raw_send_sendc sendc = raw_send_sendc::not_sendc, typename T1 , int n1, typename T2 , int n2, typename T3 , int n3> | |
| __ESIMD_API sycl::ext::intel::esimd::simd< T1, n1 > | sycl::_V1::ext::intel::esimd::raw_sends (sycl::ext::intel::esimd::simd< T1, n1 > msg_dst, sycl::ext::intel::esimd::simd< T2, n2 > msg_src0, sycl::ext::intel::esimd::simd< T3, n3 > msg_src1, uint32_t ex_desc, uint32_t msg_desc, sycl::ext::intel::esimd::simd_mask< exec_size > mask=1) |
| Raw sends. More... | |
| template<uint8_t exec_size, uint8_t sfid, uint8_t num_src0, uint8_t num_dst, raw_send_eot eot = raw_send_eot::not_eot, raw_send_sendc sendc = raw_send_sendc::not_sendc, typename T1 , int n1, typename T2 , int n2> | |
| __ESIMD_API sycl::ext::intel::esimd::simd< T1, n1 > | sycl::_V1::ext::intel::esimd::raw_send (sycl::ext::intel::esimd::simd< T1, n1 > msg_dst, sycl::ext::intel::esimd::simd< T2, n2 > msg_src0, uint32_t ex_desc, uint32_t msg_desc, sycl::ext::intel::esimd::simd_mask< exec_size > mask=1) |
| Raw send. More... | |
| template<uint8_t exec_size, uint8_t sfid, uint8_t num_src0, uint8_t num_src1, raw_send_eot eot = raw_send_eot::not_eot, raw_send_sendc sendc = raw_send_sendc::not_sendc, typename T1 , int n1, typename T2 , int n2> | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::raw_sends (sycl::ext::intel::esimd::simd< T1, n1 > msg_src0, sycl::ext::intel::esimd::simd< T2, n2 > msg_src1, uint32_t ex_desc, uint32_t msg_desc, sycl::ext::intel::esimd::simd_mask< exec_size > mask=1) |
| Raw sends. More... | |
| template<uint8_t exec_size, uint8_t sfid, uint8_t num_src0, raw_send_eot eot = raw_send_eot::not_eot, raw_send_sendc sendc = raw_send_sendc::not_sendc, typename T1 , int n1> | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::raw_send (sycl::ext::intel::esimd::simd< T1, n1 > msg_src0, uint32_t ex_desc, uint32_t msg_desc, sycl::ext::intel::esimd::simd_mask< exec_size > mask=1) |
| Raw send. More... | |
| #define __ESIMD_FP_ATOMIC_OP_TYPE_CHECK | ( | T | ) |
Definition at line 4257 of file memory.hpp.
1.9.1