|
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/simd.hpp>#include <sycl/ext/intel/esimd/simd_view.hpp>#include <sycl/half_type.hpp>#include <cstdint>
Go to the source code of this file.
Namespaces | |
| sycl | |
| ---— Error handling, matching OpenCL plugin semantics. | |
| sycl::_V1 | |
| sycl::_V1::ext | |
| sycl::_V1::ext::intel | |
| sycl::_V1::ext::intel::esimd | |
| sycl::_V1::ext::intel::esimd::detail | |
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::l3_flush_instructions = 0x2, sycl::_V1::ext::intel::esimd::l3_flush_texture_data = 0x4, sycl::_V1::ext::intel::esimd::l3_flush_constant_data = 0x8, sycl::_V1::ext::intel::esimd::l3_flush_rw_data = 0x10, sycl::_V1::ext::intel::esimd::local_barrier = 0x20, sycl::_V1::ext::intel::esimd::l1_flush_ro_data = 0x40, sycl::_V1::ext::intel::esimd::sw_barrier = 0x80 } |
| 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 Tx , int N, typename Toffset > | |
| __ESIMD_API simd< Tx, N > | sycl::_V1::ext::intel::esimd::gather (const Tx *p, simd< Toffset, N > offsets, simd_mask< N > mask=1) |
| Loads ("gathers") elements from different memory locations and returns a vector of them. More... | |
| template<typename Tx , int N, typename Toffset , typename RegionTy = region1d_t<Toffset, N, 1>> | |
| __ESIMD_API simd< Tx, N > | sycl::_V1::ext::intel::esimd::gather (const Tx *p, simd_view< Toffset, RegionTy > offsets, simd_mask< N > mask=1) |
A variation of gather API with offsets represented as simd_view object. 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 Tx , int N, typename Toffset > | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::scatter (Tx *p, simd< Toffset, N > offsets, simd< Tx, N > vals, simd_mask< N > mask=1) |
| Writes ("scatters") elements of the input vector to different memory locations. More... | |
| template<typename Tx , int N, typename Toffset , typename RegionTy = region1d_t<Toffset, N, 1>> | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::scatter (Tx *p, simd_view< Toffset, RegionTy > offsets, simd< Tx, N > vals, simd_mask< N > mask=1) |
A variation of scatter API with offsets represented as simd_view object. 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 Tx , int N, typename Flags = vector_aligned_tag, class T = detail::__raw_t<Tx>, typename = std::enable_if_t<is_simd_flag_type_v<Flags>>> | |
| __ESIMD_API simd< Tx, N > | sycl::_V1::ext::intel::esimd::block_load (const Tx *addr, Flags={}) |
| Loads a contiguous block of memory from given memory address 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>>, class T = detail::__raw_t<Tx>> | |
| __ESIMD_API simd< Tx, N > | sycl::_V1::ext::intel::esimd::block_load (AccessorTy acc, uint32_t offset, Flags={}) |
| Loads a contiguous block of memory from given accessor and offset and returns the loaded data as a vector. More... | |
| template<typename Tx , int N, class T = detail::__raw_t<Tx>> | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::block_store (Tx *p, simd< Tx, N > vals) |
| Stores elements of a vector to a contiguous block of memory at given address. More... | |
| template<typename Tx , int N, typename AccessorTy , class T = detail::__raw_t<Tx>> | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::block_store (AccessorTy acc, uint32_t offset, simd< Tx, N > vals) |
| Stores elements of a vector to a contiguous block of memory represented by an accessor and an offset within this accessor. More... | |
| template<typename T , int N, typename AccessorTy > | |
| __ESIMD_API std::enable_if_t<(sizeof(T)<=4) &&(N==1||N==8||N==16||N==32) &&!std::is_pointer< AccessorTy >::value, simd< T, N > > | sycl::_V1::ext::intel::esimd::gather (AccessorTy acc, simd< uint32_t, N > offsets, uint32_t glob_offset=0, simd_mask< N > mask=1) |
| template<typename T , int N, typename AccessorTy > | |
| __ESIMD_API std::enable_if_t<(sizeof(T)<=4) &&(N==1||N==8||N==16||N==32) &&!std::is_pointer< AccessorTy >::value > | sycl::_V1::ext::intel::esimd::scatter (AccessorTy acc, simd< uint32_t, N > offsets, simd< T, N > vals, uint32_t glob_offset=0, simd_mask< N > mask=1) |
| template<typename T , typename AccessorTy > | |
| __ESIMD_API T | sycl::_V1::ext::intel::esimd::scalar_load (AccessorTy acc, uint32_t 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, uint32_t 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) |
| template<rgba_channel_mask RGBAMask = rgba_channel_mask::ABGR, typename T , int N, typename Toffset , typename RegionTy = region1d_t<Toffset, N, 1>> | |
| __ESIMD_API simd< T, N *get_num_channels_enabled(RGBAMask)> | sycl::_V1::ext::intel::esimd::gather_rgba (const T *p, simd_view< Toffset, RegionTy > 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<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::gather_rgba (const T *p, simd< uint32_t, N > offsets, simd_mask< N > mask=1) |
| 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) |
| template<rgba_channel_mask RGBAMask = rgba_channel_mask::ABGR, typename T , int N, typename Toffset , typename RegionTy = region1d_t<Toffset, N, 1>> | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::scatter_rgba (T *p, simd_view< Toffset, RegionTy > 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 &&!std::is_pointer_v< AccessorT >), 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) |
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 &&!std::is_pointer_v< AccessorT > > | 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) |
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> | |
| 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<atomic_op Op, typename Tx , int N, typename Toffset > | |
| __ESIMD_API simd< Tx, N > | sycl::_V1::ext::intel::esimd::atomic_update (Tx *p, simd< Toffset, N > offset, simd< Tx, N > src0, simd_mask< N > mask) |
| Single-argument variant of the atomic update operation. More... | |
| template<atomic_op Op, typename Tx , int N, typename Toffset , typename RegionTy = region1d_t<Toffset, N, 1>> | |
| __ESIMD_API simd< Tx, N > | sycl::_V1::ext::intel::esimd::atomic_update (Tx *p, simd_view< Toffset, RegionTy > offsets, simd< Tx, N > src0, simd_mask< N > mask) |
A variation of atomic_update API with offsets represented as simd_view object. 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 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 Tx , int N, typename Toffset > | |
| __ESIMD_API simd< Tx, N > | sycl::_V1::ext::intel::esimd::atomic_update (Tx *p, simd< Toffset, N > offset, simd_mask< N > mask) |
| No-argument variant of the atomic update operation. More... | |
| template<atomic_op Op, typename Tx , int N, typename Toffset , typename RegionTy = region1d_t<Toffset, N, 1>> | |
| __ESIMD_API simd< Tx, N > | sycl::_V1::ext::intel::esimd::atomic_update (Tx *p, simd_view< Toffset, RegionTy > offsets, simd_mask< N > mask=1) |
A variation of atomic_update API with offsets represented as simd_view object. 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 offset, simd_mask< N > mask=1) |
A variation of atomic_update API with offset represented as scalar. More... | |
| template<atomic_op Op, typename Tx , int N, typename Toffset > | |
| __ESIMD_API simd< Tx, N > | sycl::_V1::ext::intel::esimd::atomic_update (Tx *p, simd< Toffset, N > offset, simd< Tx, N > src0, simd< Tx, N > src1, simd_mask< N > mask) |
| template<atomic_op Op, typename Tx , int N, typename Toffset , typename RegionTy = region1d_t<Toffset, N, 1>> | |
| __ESIMD_API simd< Tx, N > | sycl::_V1::ext::intel::esimd::atomic_update (Tx *p, simd_view< Toffset, RegionTy > offsets, simd< Tx, N > src0, simd< Tx, N > src1, simd_mask< N > mask) |
A variation of atomic_update API with offsets represented as simd_view object. 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 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<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) |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::barrier () |
| Generic work-group barrier. 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> | |
| __ESIMD_API std::enable_if_t<(N==1||N==8||N==16||N==32), simd< T, N > > | sycl::_V1::ext::intel::esimd::slm_gather (simd< uint32_t, N > offsets, simd_mask< N > mask=1) |
| Gather operation over the Shared Local Memory. 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> | |
| __ESIMD_API std::enable_if_t<(N==1||N==8||N==16||N==32) &&(sizeof(T)<=4)> | sycl::_V1::ext::intel::esimd::slm_scatter (simd< uint32_t, N > offsets, simd< T, N > vals, simd_mask< N > mask=1) |
| Scatter operation over the Shared Local Memory. 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> | |
| __ESIMD_API simd< T, N > | sycl::_V1::ext::intel::esimd::slm_block_load (uint32_t offset) |
| Loads a contiguous block of memory from the SLM at given offset and returns the loaded data as a vector. More... | |
| template<typename T , int N> | |
| __ESIMD_API void | sycl::_V1::ext::intel::esimd::slm_block_store (uint32_t offset, simd< T, N > vals) |
| Stores elements of a vector to a contiguous block of SLM at given offset. More... | |
| template<atomic_op Op, typename Tx , int N, class T = detail::__raw_t<Tx>> | |
| __ESIMD_API simd< Tx, N > | sycl::_V1::ext::intel::esimd::slm_atomic_update (simd< uint32_t, N > offsets, simd_mask< N > mask) |
| Atomic update operation performed on SLM. More... | |
| template<atomic_op Op, typename Tx , int N, class T = detail::__raw_t<Tx>> | |
| __ESIMD_API simd< Tx, N > | sycl::_V1::ext::intel::esimd::slm_atomic_update (simd< uint32_t, N > offsets, simd< Tx, N > src0, simd_mask< N > mask) |
| Atomic update operation performed on SLM. More... | |
| template<atomic_op Op, typename Tx , int N, class T = detail::__raw_t<Tx>> | |
| __ESIMD_API simd< Tx, N > | sycl::_V1::ext::intel::esimd::slm_atomic_update (simd< uint32_t, N > offsets, simd< Tx, N > src0, simd< Tx, N > src1, simd_mask< N > mask) |
| Atomic update operation performed on SLM. 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... | |
1.8.17