|
DPC++ Runtime
Runtime libraries for oneAPI DPC++
|
|
|
DPC++ Runtime
Runtime libraries for oneAPI DPC++
|
|
Namespaces | |
| detail | |
| native | |
| xmx | |
Classes | |
| struct | element_aligned_tag |
element_aligned_tag type. More... | |
| struct | is_simd_flag_type |
| Checks if type is a simd load/store flag. More... | |
| struct | is_simd_flag_type< detail::dqword_element_aligned_tag > |
| struct | is_simd_flag_type< element_aligned_tag > |
| struct | is_simd_flag_type< overaligned_tag< N > > |
| struct | is_simd_flag_type< vector_aligned_tag > |
| struct | overaligned_tag |
overaligned_tag type. More... | |
| struct | saturation_off_tag |
| This type tag represents "saturation off" behavior. More... | |
| struct | saturation_on_tag |
| Gen hardware supports applying saturation to results of certain operations. More... | |
| class | simd |
| The main simd vector class. More... | |
| class | simd_view |
| This class represents a reference to a sub-region of a base simd object. More... | |
| class | simd_view< BaseTy, region1d_scalar_t< ViewedElemT > > |
| This is a specialization of simd_view class with a single element. More... | |
| class | simd_view< BaseTy, std::pair< region1d_scalar_t< ViewedElemT >, NestedRegion > > |
| This is a specialization of nested simd_view class with a single element. More... | |
| class | slm_allocator |
| RAII-style class used to implement "semi-dynamic" SLM allocation. More... | |
| struct | vector_aligned_tag |
vector_aligned_tag type. More... | |
Typedefs | |
| using | uchar = unsigned char |
| using | ushort = unsigned short |
| using | uint = unsigned int |
| using | SurfaceIndex = unsigned int |
| Surface index type. More... | |
| template<int N> | |
| using | mask_type_t = detail::simd_mask_storage_t< N > |
| template<int N> | |
| using | simd_mask = detail::simd_mask_type< N > |
Represents a simd mask os size N. More... | |
Enumerations | |
| enum | rgba_channel : uint8_t { rgba_channel::R, rgba_channel::G, rgba_channel::B, rgba_channel::A } |
| Represents a pixel's channel. More... | |
| enum | rgba_channel_mask : uint8_t { rgba_channel_mask::R = detail::chR, rgba_channel_mask::G = detail::chG, rgba_channel_mask::GR = detail::chG | detail::chR, rgba_channel_mask::B = detail::chB, rgba_channel_mask::BR = detail::chB | detail::chR, rgba_channel_mask::BG = detail::chB | detail::chG, rgba_channel_mask::BGR = detail::chB | detail::chG | detail::chR, rgba_channel_mask::A = detail::chA, rgba_channel_mask::AR = detail::chA | detail::chR, rgba_channel_mask::AG = detail::chA | detail::chG, rgba_channel_mask::AGR = detail::chA | detail::chG | detail::chR, rgba_channel_mask::AB = detail::chA | detail::chB, rgba_channel_mask::ABR = detail::chA | detail::chB | detail::chR, rgba_channel_mask::ABG = detail::chA | detail::chB | detail::chG, rgba_channel_mask::ABGR = detail::chA | detail::chB | detail::chG | detail::chR } |
| Represents a pixel's channel mask - all possible combinations of enabled channels. More... | |
| enum | atomic_op : uint8_t { atomic_op::add = 0x0, atomic_op::sub = 0x1, atomic_op::inc = 0x2, atomic_op::dec = 0x3, atomic_op::umin = 0x4, atomic_op::__SYCL_DEPRECATED =("use umin") = umin, atomic_op::umax = 0x5, atomic_op::__SYCL_DEPRECATED =("use smax") = umax, atomic_op::xchg = 0x6, atomic_op::cmpxchg = 0x7, atomic_op::bit_and = 0x8, atomic_op::bit_or = 0x9, atomic_op::bit_xor = 0xa, atomic_op::smin = 0xb, atomic_op::__SYCL_DEPRECATED =("use smin") = smin, atomic_op::smax = 0xc, atomic_op::__SYCL_DEPRECATED =("use smax") = 0xc, atomic_op::__SYCL_DEPRECATED =("fmax" __ESIMD_USM_DWORD_ATOMIC_TO_LSC) = 0x10, atomic_op::__SYCL_DEPRECATED =("fmin" __ESIMD_USM_DWORD_ATOMIC_TO_LSC) = 0x11, atomic_op::fcmpxchg = 0x12, atomic_op::__SYCL_DEPRECATED =("fcmpwr" __ESIMD_USM_DWORD_ATOMIC_TO_LSC) = fcmpxchg, atomic_op::__SYCL_DEPRECATED =("fadd" __ESIMD_USM_DWORD_ATOMIC_TO_LSC) = 0x13, atomic_op::__SYCL_DEPRECATED =("fsub" __ESIMD_USM_DWORD_ATOMIC_TO_LSC) = 0x14, atomic_op::load = 0x15, atomic_op::store = 0x16, atomic_op::predec = 0xff } |
| Represents an atomic operation. More... | |
| enum | fence_mask : uint8_t { global_coherent_fence = 0x1, l3_flush_instructions = 0x2, l3_flush_texture_data = 0x4, l3_flush_constant_data = 0x8, l3_flush_rw_data = 0x10, local_barrier = 0x20, l1_flush_ro_data = 0x40, sw_barrier = 0x80 } |
| Represetns a bit mask to control behavior of esimd::fence. More... | |
Functions | |
| template<class T , int N> | |
| __ESIMD_API simd< T, N > | merge (simd< T, N > a, simd< T, N > b, simd_mask< N > m) |
| "Merges" elements of the input simd object according to the merge mask. More... | |
| template<int N> | |
| __ESIMD_API simd_mask< N > | merge (simd_mask< N > a, simd_mask< N > b, simd_mask< N > m) |
| "Merges" elements of the input masks according to the merge mask. More... | |
| template<class BaseT1 , class BaseT2 , class RegionT1 , class RegionT2 , class = std::enable_if_t< (shape_type<RegionT1>::length == shape_type<RegionT2>::length) && std::is_same_v<detail::element_type_t<BaseT1>, detail::element_type_t<BaseT2>>>> | |
| __ESIMD_API auto | merge (simd_view< BaseT1, RegionT1 > v1, simd_view< BaseT2, RegionT2 > v2, simd_mask< shape_type< RegionT1 >::length > m) |
| "Merges" elements of vectors referenced by the input views. More... | |
| constexpr int | is_channel_enabled (rgba_channel_mask M, rgba_channel Ch) |
| constexpr int | get_num_channels_enabled (rgba_channel_mask M) |
| template<typename T0 , typename T1 , int SZ> | |
| __ESIMD_API std::enable_if_t<!detail::is_generic_floating_point_v< T0 >||std::is_same_v< T1, T0 >, simd< T0, SZ > > | saturate (simd< T1, SZ > src) |
Conversion of input vector elements of type T1 into vector of elements of type T0 with saturation. More... | |
| template<typename TRes , typename TArg , int SZ> | |
| ESIMD_DETAIL __ESIMD_API std::enable_if_t< !std::is_same< std::remove_const_t< TRes >, std::remove_const_t< TArg > >::value, simd< TRes, SZ > > | abs (simd< TArg, SZ > src0) |
| Get absolute value (vector version) More... | |
| template<typename TRes , typename TArg > | |
| __ESIMD_API std::enable_if_t<!std::is_same< std::remove_const_t< TRes >, std::remove_const_t< TArg > >::value &&detail::is_esimd_scalar< TRes >::value &&detail::is_esimd_scalar< TArg >::value, std::remove_const_t< TRes > > | abs (TArg src0) |
| Get absolute value (scalar version) More... | |
| template<typename T1 , int SZ> | |
| __ESIMD_API simd< T1, SZ > | abs (simd< T1, SZ > src0) |
| Get absolute value (vector version). More... | |
| template<typename T1 > | |
| __ESIMD_API std::enable_if_t< detail::is_esimd_scalar< T1 >::value, std::remove_const_t< T1 > > | abs (T1 src0) |
| Get absolute value (scalar version). More... | |
| template<typename T , int SZ, class Sat = saturation_off_tag> | |
| __ESIMD_API simd< T, SZ > | max (simd< T, SZ > src0, simd< T, SZ > src1, Sat sat={}) |
| Selects component-wise the maximum of the two vectors. More... | |
| template<typename T , int SZ, class Sat = saturation_off_tag> | |
| __ESIMD_API std::enable_if_t< detail::is_esimd_scalar< T >::value, simd< T, SZ > > | max (simd< T, SZ > src0, T src1, Sat sat={}) |
| Selects maximums for each element of the input vector and a scalar. More... | |
| template<typename T , int SZ, class Sat = saturation_off_tag> | |
| __ESIMD_API std::enable_if_t< detail::is_esimd_scalar< T >::value, simd< T, SZ > > | max (T src0, simd< T, SZ > src1, Sat sat={}) |
| Selects maximums for each element of the input scalar and a vector. More... | |
| template<typename T , class Sat = saturation_off_tag> | |
| ESIMD_NODEBUG ESIMD_INLINE std::enable_if_t< detail::is_esimd_scalar< T >::value, T > | max (T src0, T src1, Sat sat={}) |
| Selects maximum between two scalar values. More... | |
| template<typename T , int SZ, class Sat = saturation_off_tag> | |
| __ESIMD_API simd< T, SZ > | min (simd< T, SZ > src0, simd< T, SZ > src1, Sat sat={}) |
| Selects component-wise the minimum of the two vectors. More... | |
| template<typename T , int SZ, class Sat = saturation_off_tag> | |
| __ESIMD_API std::enable_if_t< detail::is_esimd_scalar< T >::value, simd< T, SZ > > | min (simd< T, SZ > src0, T src1, Sat sat={}) |
| Selects minimums for each element of the input vector and a scalar. More... | |
| template<typename T , int SZ, class Sat = saturation_off_tag> | |
| __ESIMD_API std::enable_if_t< detail::is_esimd_scalar< T >::value, simd< T, SZ > > | min (T src0, simd< T, SZ > src1, Sat sat={}) |
| Selects minimums for each element of the input scalar and a vector. More... | |
| template<typename T , class Sat = saturation_off_tag> | |
| ESIMD_NODEBUG ESIMD_INLINE std::enable_if_t< detail::is_esimd_scalar< T >::value, T > | min (T src0, T src1, Sat sat={}) |
| Selects minimum between two scalar values. More... | |
| template<class T , int N, class Sat = saturation_off_tag, class = std::enable_if_t< detail::is_generic_floating_point_v<T> && (sizeof(T) <= 4) >> | |
| __ESIMD_API simd< T, N > | inv (simd< T, N > src, Sat sat={}) |
| Inversion - calculates (1/x). More... | |
| template<typename T , class Sat = saturation_off_tag, class = std::enable_if_t< __ESIMD_EMATH_COND >> | |
| __ESIMD_API T | inv (T src, Sat sat={}) |
| Scalar version. More... | |
| template<class T , int N, class Sat = saturation_off_tag, class = std::enable_if_t< detail::is_generic_floating_point_v<T> && (sizeof(T) <= 4) >> | |
| __ESIMD_API simd< T, N > | log2 (simd< T, N > src, Sat sat={}) |
| Logarithm base 2. More... | |
| template<typename T , class Sat = saturation_off_tag, class = std::enable_if_t< __ESIMD_EMATH_COND >> | |
| __ESIMD_API T | log2 (T src, Sat sat={}) |
| Scalar version. More... | |
| template<class T , int N, class Sat = saturation_off_tag, class = std::enable_if_t< detail::is_generic_floating_point_v<T> && (sizeof(T) <= 4) >> | |
| __ESIMD_API simd< T, N > | exp2 (simd< T, N > src, Sat sat={}) |
| Exponent base 2. More... | |
| template<typename T , class Sat = saturation_off_tag, class = std::enable_if_t< __ESIMD_EMATH_COND >> | |
| __ESIMD_API T | exp2 (T src, Sat sat={}) |
| Scalar version. More... | |
| template<class T , int N, class Sat = saturation_off_tag, class = std::enable_if_t< detail::is_generic_floating_point_v<T> && (sizeof(T) <= 4) >> | |
| __ESIMD_API simd< T, N > | sqrt (simd< T, N > src, Sat sat={}) |
| Square root. More... | |
| template<typename T , class Sat = saturation_off_tag, class = std::enable_if_t< __ESIMD_EMATH_COND >> | |
| __ESIMD_API T | sqrt (T src, Sat sat={}) |
| Scalar version. More... | |
| template<class T , int N, class Sat = saturation_off_tag, class = std::enable_if_t< detail::is_generic_floating_point_v<T> && (sizeof(T) >= 4) >> | |
| __ESIMD_API simd< T, N > | sqrt_ieee (simd< T, N > src, Sat sat={}) |
IEEE754-compliant square root. Supports float and double. More... | |
| template<typename T , class Sat = saturation_off_tag, class = std::enable_if_t< __ESIMD_EMATH_IEEE_COND >> | |
| __ESIMD_API T | sqrt_ieee (T src, Sat sat={}) |
| Scalar version. More... | |
| template<class T , int N, class Sat = saturation_off_tag, class = std::enable_if_t< detail::is_generic_floating_point_v<T> && (sizeof(T) <= 4) >> | |
| __ESIMD_API simd< T, N > | rsqrt (simd< T, N > src, Sat sat={}) |
Square root reciprocal - calculates 1/sqrt(x). More... | |
| template<typename T , class Sat = saturation_off_tag, class = std::enable_if_t< __ESIMD_EMATH_COND >> | |
| __ESIMD_API T | rsqrt (T src, Sat sat={}) |
| Scalar version. More... | |
| template<class T , int N, class Sat = saturation_off_tag, class = std::enable_if_t< detail::is_generic_floating_point_v<T> && (sizeof(T) <= 4) >> | |
| __ESIMD_API simd< T, N > | sin (simd< T, N > src, Sat sat={}) |
| Sine. More... | |
| template<typename T , class Sat = saturation_off_tag, class = std::enable_if_t< __ESIMD_EMATH_COND >> | |
| __ESIMD_API T | sin (T src, Sat sat={}) |
| Scalar version. More... | |
| template<class T , int N, class Sat = saturation_off_tag, class = std::enable_if_t< detail::is_generic_floating_point_v<T> && (sizeof(T) <= 4) >> | |
| __ESIMD_API simd< T, N > | cos (simd< T, N > src, Sat sat={}) |
| Cosine. More... | |
| template<typename T , class Sat = saturation_off_tag, class = std::enable_if_t< __ESIMD_EMATH_COND >> | |
| __ESIMD_API T | cos (T src, Sat sat={}) |
| Scalar version. More... | |
| template<class T , int N, class U , class Sat = saturation_off_tag, class = std::enable_if_t< detail::is_generic_floating_point_v<T> && (sizeof(T) <= 4) >> | |
| __ESIMD_API simd< T, N > | pow (simd< T, N > src0, simd< U, N > src1, Sat sat={}) |
Power - calculates src0 in power of src1. More... | |
| template<class T , int N, class U , class Sat = saturation_off_tag, class = std::enable_if_t< __ESIMD_EMATH_COND >> | |
| __ESIMD_API simd< T, N > | pow (simd< T, N > src0, U src1, Sat sat={}) |
| (vector, scalar) version. More... | |
| template<class T , class U , class Sat = saturation_off_tag, class = std::enable_if_t< __ESIMD_EMATH_COND >> | |
| __ESIMD_API T | pow (T src0, U src1, Sat sat={}) |
| (scalar, scalar) version. More... | |
| template<class T , int N, class U , class Sat = saturation_off_tag, class = std::enable_if_t< detail::is_generic_floating_point_v<T> && (sizeof(T) >= 4) >> | |
| __ESIMD_API simd< T, N > | div_ieee (simd< T, N > src0, simd< U, N > src1, Sat sat={}) |
IEEE754-compliant floating-point division. Supports float and double. More... | |
| template<class T , int N, class U , class Sat = saturation_off_tag, class = std::enable_if_t< __ESIMD_EMATH_IEEE_COND >> | |
| __ESIMD_API simd< T, N > | div_ieee (simd< T, N > src0, U src1, Sat sat={}) |
| (vector, scalar) version. More... | |
| template<class T , class U , class Sat = saturation_off_tag, class = std::enable_if_t< __ESIMD_EMATH_IEEE_COND >> | |
| __ESIMD_API T | div_ieee (T src0, U src1, Sat sat={}) |
| (scalar, scalar) version. More... | |
| template<class T , int SZ, class Sat = saturation_off_tag> | |
| ESIMD_DETAIL ESIMD_NODEBUG ESIMD_INLINE simd< T, SZ > | log (simd< T, SZ > src0, Sat sat={}) |
| Computes the natural logarithm of the given argument. More... | |
| template<class T , class Sat = saturation_off_tag> | |
| ESIMD_NODEBUG ESIMD_INLINE T | log (T src0, Sat sat={}) |
| template<class T , int SZ, class Sat = saturation_off_tag> | |
| ESIMD_NODEBUG ESIMD_INLINE simd< T, SZ > | exp (simd< T, SZ > src0, Sat sat={}) |
| Computes e raised to the power of the given argument. More... | |
| template<class T , class Sat = saturation_off_tag> | |
| ESIMD_NODEBUG ESIMD_INLINE T | exp (T src0, Sat sat={}) |
| template<typename T , int SZ, class Sat = sycl::ext::intel::esimd ::saturation_off_tag> | |
| __ESIMD_API __ESIMD_NS::simd< T, SZ > | rndd (__ESIMD_NS::simd< float, SZ > src0, Sat sat={}) |
Round-down (also known as floor). More... | |
| template<typename T , class Sat = __ESIMD_NS::saturation_off_tag> | |
| __ESIMD_API T | rndd (float src0, Sat sat={}) |
| Scalar version. More... | |
| template<typename T , int SZ, class Sat = sycl::ext::intel::esimd ::saturation_off_tag> | |
| __ESIMD_API __ESIMD_NS::simd< T, SZ > | rndu (__ESIMD_NS::simd< float, SZ > src0, Sat sat={}) |
Round-up (also known as ceil). More... | |
| template<typename T , class Sat = __ESIMD_NS::saturation_off_tag> | |
| __ESIMD_API T | rndu (float src0, Sat sat={}) |
| Scalar version. More... | |
| template<typename T , int SZ, class Sat = sycl::ext::intel::esimd ::saturation_off_tag> | |
| __ESIMD_API __ESIMD_NS::simd< T, SZ > | rnde (__ESIMD_NS::simd< float, SZ > src0, Sat sat={}) |
Round-to-even (also known as round). More... | |
| template<typename T , class Sat = __ESIMD_NS::saturation_off_tag> | |
| __ESIMD_API T | rnde (float src0, Sat sat={}) |
| Scalar version. More... | |
| template<typename T , int SZ, class Sat = sycl::ext::intel::esimd ::saturation_off_tag> | |
| __ESIMD_API __ESIMD_NS::simd< T, SZ > | rndz (__ESIMD_NS::simd< float, SZ > src0, Sat sat={}) |
Round-to-zero (also known as trunc). More... | |
| template<typename T , class Sat = __ESIMD_NS::saturation_off_tag> | |
| __ESIMD_API T | rndz (float src0, Sat sat={}) |
| Scalar version. More... | |
| template<typename RT , int SZ, class Sat = sycl::ext::intel::esimd::saturation_off_tag> | |
| ESIMD_INLINE sycl::ext::intel::esimd::simd< RT, SZ > | floor (const sycl::ext::intel::esimd::simd< float, SZ > src0, Sat sat={}) |
"Floor" operation, vector version - alias of rndd. More... | |
| template<typename RT , class Sat = sycl::ext::intel::esimd::saturation_off_tag> | |
| ESIMD_INLINE RT | floor (float src0, Sat sat={}) |
"Floor" operation, scalar version - alias of rndd. More... | |
| template<typename RT , int SZ, class Sat = sycl::ext::intel::esimd::saturation_off_tag> | |
| ESIMD_INLINE sycl::ext::intel::esimd::simd< RT, SZ > | ceil (const sycl::ext::intel::esimd::simd< float, SZ > src0, Sat sat={}) |
"Ceiling" operation, vector version - alias of rndu. More... | |
| template<typename RT , class Sat = sycl::ext::intel::esimd::saturation_off_tag> | |
| ESIMD_INLINE RT | ceil (float src0, Sat sat={}) |
"Ceiling" operation, scalar version - alias of rndu. More... | |
| template<typename RT , int SZ, class Sat = sycl::ext::intel::esimd::saturation_off_tag> | |
| __ESIMD_API sycl::ext::intel::esimd::simd< RT, SZ > | trunc (const sycl::ext::intel::esimd::simd< float, SZ > &src0, Sat sat={}) |
| Round to integral value using the round to zero rounding mode (vector version). More... | |
| template<typename RT , class Sat = sycl::ext::intel::esimd::saturation_off_tag> | |
| __ESIMD_API RT | trunc (float src0, Sat sat={}) |
| Round to integral value using the round to zero rounding mode (scalar version). More... | |
| template<int N> | |
| ESIMD_NODEBUG ESIMD_INLINE std::enable_if_t<(N==8||N==16||N==32), uint > | pack_mask (simd_mask< N > src0) |
| Pack a simd_mask into a single unsigned 32-bit integer value. More... | |
| template<int N> | |
| ESIMD_NODEBUG ESIMD_INLINE std::enable_if_t<(N==8||N==16||N==32), simd_mask< N > > | unpack_mask (uint src0) |
| Unpack an unsigned 32-bit integer value into a simd_mask. More... | |
| template<int N> | |
| __ESIMD_API std::enable_if_t<(N !=8 &&N !=16 &&N< 32), uint > | pack_mask (simd_mask< N > src0) |
pack_mask specialization when the number of elements N is not 8, 16 or 32. More... | |
| template<typename T , int N> | |
| __ESIMD_API std::enable_if_t<(std::is_same_v< T, ushort >||std::is_same_v< T, uint >)&&(N > 0 &&N<=32), uint > | ballot (simd< T, N > mask) |
| Compare source vector elements against zero and return a bitfield combining the comparison result. More... | |
| template<typename T , int N> | |
| ESIMD_NODEBUG ESIMD_INLINE std::enable_if_t< std::is_integral< T >::value &&(sizeof(T)<=4), simd< uint32_t, N > > | cbit (simd< T, N > src) |
| Count number of bits set in the source operand per element. More... | |
| template<typename T > | |
| __ESIMD_API std::enable_if_t< std::is_integral< T >::value &&(sizeof(T)<=4), uint32_t > | cbit (T src) |
Scalar version of cbit - both input and output are scalars rather than vectors. More... | |
| template<typename BaseTy , typename RegionTy > | |
| __ESIMD_API std::enable_if_t< std::is_integral< typename simd_view< BaseTy, RegionTy >::element_type >::value &&(sizeof(typename simd_view< BaseTy, RegionTy >::element_type)<=4) &&(simd_view< BaseTy, RegionTy >::length==1), uint32_t > | cbit (simd_view< BaseTy, RegionTy > src) |
Scalar version of cbit, that takes simd_view object as an argument, e.g. More... | |
| template<typename T , int N> | |
| __ESIMD_API std::enable_if_t< std::is_integral< T >::value &&(sizeof(T)==4), simd< T, N > > | fbl (simd< T, N > src) |
| Find the per element number of the first bit set in the source operand starting from the least significant bit. More... | |
| template<typename T > | |
| __ESIMD_API std::enable_if_t< std::is_integral< T >::value &&(sizeof(T)==4), T > | fbl (T src) |
Scalar version of fbl - both input and output are scalars rather than vectors. More... | |
| template<typename BaseTy , typename RegionTy > | |
| __ESIMD_API std::enable_if_t< std::is_integral< typename simd_view< BaseTy, RegionTy >::element_type >::value &&(sizeof(typename simd_view< BaseTy, RegionTy >::element_type)==4) &&(simd_view< BaseTy, RegionTy >::length==1), typename simd_view< BaseTy, RegionTy >::element_type > | fbl (simd_view< BaseTy, RegionTy > src) |
Scalar version of fbl, that takes simd_view object as an argument, e.g. More... | |
| template<typename T , int N> | |
| __ESIMD_API std::enable_if_t< std::is_integral< T >::value &&std::is_signed< T >::value &&(sizeof(T)==4), simd< T, N > > | fbh (simd< T, N > src) |
| Find the per element number of the first bit set in the source operand starting from the most significant bit (sign bit is skipped). More... | |
| template<typename T , int N> | |
| __ESIMD_API std::enable_if_t< std::is_integral< T >::value &&!std::is_signed< T >::value &&(sizeof(T)==4), simd< T, N > > | fbh (simd< T, N > src) |
| Find the per element number of the first bit set in the source operand starting from the most significant bit (sign bit is counted). More... | |
| template<typename T > | |
| __ESIMD_API std::enable_if_t< std::is_integral< T >::value &&(sizeof(T)==4), T > | fbh (T src) |
Scalar version of fbh - both input and output are scalars rather than vectors. More... | |
| template<typename BaseTy , typename RegionTy > | |
| __ESIMD_API std::enable_if_t< std::is_integral< typename simd_view< BaseTy, RegionTy >::element_type >::value &&(sizeof(typename simd_view< BaseTy, RegionTy >::element_type)==4) &&(simd_view< BaseTy, RegionTy >::length==1), typename simd_view< BaseTy, RegionTy >::element_type > | fbh (simd_view< BaseTy, RegionTy > src) |
Scalar version of fbh, that takes simd_view object as an argument, e.g. More... | |
| template<typename T1 , typename T2 , typename T3 , typename T4 , int N, class Sat = saturation_off_tag> | |
| __ESIMD_API std::enable_if_t< detail::is_dword_type< T1 >::value &&detail::is_dword_type< T2 >::value &&detail::is_dword_type< T3 >::value &&detail::is_dword_type< T4 >::value, simd< T1, N > > | dp4a (simd< T2, N > src0, simd< T3, N > src1, simd< T4, N > src2, Sat sat={}) |
| DP4A. More... | |
| template<typename T0 , typename T1 , int SZ> | |
| ESIMD_INLINE ESIMD_NODEBUG T0 | hmax (simd< T1, SZ > v) |
| ESIMD_DETAIL. More... | |
| template<typename T0 , typename T1 , int SZ> | |
| ESIMD_INLINE ESIMD_NODEBUG T0 | hmin (simd< T1, SZ > v) |
| Performs 'minimum' operation reduction over elements of the input vector, that is, returns the minimal vector element. More... | |
| template<typename T0 , typename T1 , int SZ, typename BinaryOperation > | |
| ESIMD_INLINE ESIMD_NODEBUG T0 | reduce (simd< T1, SZ > v, BinaryOperation op) |
| Performs reduction over elements of the input vector. More... | |
| template<typename AccessorTy > | |
| __ESIMD_API SurfaceIndex | 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 > | 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 > | 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 > > | 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 | 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 | 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 > | 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 > | 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 > | 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 | 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 | 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 > > | 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 > | 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 | scalar_load (AccessorTy acc, uint32_t offset) |
| Load a scalar value from an accessor. More... | |
| template<typename T , typename AccessorTy > | |
| __ESIMD_API void | 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)> | 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)> | 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)> > | 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)> > | gather_rgba (const T *p, simd< uint32_t, N > offsets, simd_mask< N > mask=1) |
| template<rgba_channel_mask RGBAMask = rgba_channel_mask::ABGR, typename T , int N, typename Toffset > | |
| __ESIMD_API void | 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 | 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 > | 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 > | 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)> > | 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 > > | 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<atomic_op Op, typename Tx , int N, typename Toffset > | |
| __ESIMD_API simd< Tx, N > | 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 > | 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 > > | 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 > | 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 > | 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 > > | 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 > | 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 > | 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 > > | 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 | fence () |
| esimd::fence sets the memory read/write order. More... | |
| __ESIMD_API void | fence (fence_mask cntl) |
| __ESIMD_API void | barrier () |
| Generic work-group barrier. More... | |
| template<uint32_t SLMSize> | |
| __ESIMD_API void | slm_init () |
| Declare per-work-group slm size. More... | |
| __ESIMD_API void | 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 > > | 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 | 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)> | 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 | 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)> > | 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)> | 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 > | 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 | 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 > | 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 > | 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 > | 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 > | 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 | media_block_store (AccessorTy acc, unsigned x, unsigned y, simd< T, m *N > vals) |
| Media block store. More... | |
| template<typename To , typename From , int N> | |
| ESIMD_INLINE simd< To, N > | convert (const simd< From, N > &val) |
Covert from a simd object with element type From to a simd object with element type To. More... | |
| template<native::lsc::atomic_op Op, typename T , int N, typename Toffset > | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==0, simd< T, N > > | atomic_update (T *p, simd< Toffset, N > offset, simd_mask< N > mask) |
LSC version of no argument variant of the atomic_update - accepts native::lsc::atomic_op instead of atomic_op as atomic operation template argument. More... | |
| template<native::lsc::atomic_op Op, typename T , int N, typename Toffset , typename RegionTy = sycl::ext::intel::esimd::region1d_t<Toffset, N, 1>> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==0, simd< T, N > > | atomic_update (T *p, simd_view< Toffset, RegionTy > offsets, simd_mask< N > mask=1) |
| template<native::lsc::atomic_op Op, typename T , int N, typename Toffset > | |
| __ESIMD_API std::enable_if_t< std::is_integral_v< Toffset > &&__ESIMD_DNS::get_num_args< Op >)==0, simd< T, N > > | atomic_update (T *p, Toffset offset, simd_mask< N > mask=1) |
| template<native::lsc::atomic_op Op, typename T , int N, typename Toffset > | |
| __ESIMD_API __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==1, simd< T, N > > | atomic_update (T *p, simd< Toffset, N > offset, simd< T, N > src0, simd_mask< N > mask) |
| LSC version of the single-argument atomic update. More... | |
| template<native::lsc::atomic_op Op, typename T , int N, typename Toffset , typename RegionTy = sycl::ext::intel::esimd::region1d_t<Toffset, N, 1>> | |
| __ESIMD_API __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==1, simd< T, N > > | atomic_update (T *p, simd_view< Toffset, RegionTy > offsets, simd< T, N > src0, simd_mask< N > mask=1) |
| template<native::lsc::atomic_op Op, typename T , int N, typename Toffset > | |
| __ESIMD_API std::enable_if_t< std::is_integral_v< Toffset > &&__ESIMD_DNS::get_num_args< Op >)==1, simd< T, N > > | atomic_update (T *p, Toffset offset, simd< T, N > src0, simd_mask< N > mask=1) |
| template<native::lsc::atomic_op Op, typename T , int N, typename Toffset > | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==2, simd< T, N > > | atomic_update (T *p, simd< Toffset, N > offset, simd< T, N > src0, simd< T, N > src1, simd_mask< N > mask) |
| LSC version of the two-argument atomic update. More... | |
| template<native::lsc::atomic_op Op, typename T , int N, typename Toffset , typename RegionTy = sycl::ext::intel::esimd::region1d_t<Toffset, N, 1>> | |
| __ESIMD_API std::enable_if_t< __ESIMD_DNS::get_num_args< Op >)==2, simd< T, N > > | atomic_update (T *p, simd_view< Toffset, RegionTy > offsets, simd< T, N > src0, simd< T, N > src1, simd_mask< N > mask=1) |
| template<native::lsc::atomic_op Op, typename T , int N, typename Toffset > | |
| __ESIMD_API std::enable_if_t< std::is_integral_v< Toffset > &&__ESIMD_DNS::get_num_args< Op >)==2, sycl::ext::intel::esimd::simd< T, N > > | atomic_update (T *p, Toffset offset, simd< T, N > src0, simd< T, N > src1, simd_mask< N > mask=1) |
Variables | |
| static constexpr saturation_off_tag | saturation_off {} |
| Type tag object representing "saturation off" behavior. More... | |
| static constexpr saturation_on_tag | saturation_on {} |
| Type tag object representing "saturation on" behavior. More... | |
| constexpr element_aligned_tag | element_aligned = {} |
| constexpr vector_aligned_tag | vector_aligned = {} |
| template<unsigned N> | |
| constexpr overaligned_tag< N > | overaligned = {} |
| template<typename T > | |
| static constexpr bool | is_simd_flag_type_v = is_simd_flag_type<T>::value |
| Checks if given type is a simd load/store flag. More... | |
| using sycl::_V1::ext::intel::esimd::mask_type_t = typedef detail::simd_mask_storage_t<N> |
| __ESIMD_API std::enable_if_t<__ESIMD_DNS::get_num_args<Op>) == 2, simd<T, N> > sycl::_V1::ext::intel::esimd::atomic_update | ( | T * | p, |
| simd< Toffset, N > | offset, | ||
| simd< T, N > | src0, | ||
| simd< T, N > | src1, | ||
| simd_mask< N > | mask | ||
| ) |
LSC version of the two-argument atomic update.
Definition at line 3159 of file memory.hpp.
| __ESIMD_API __ESIMD_API std::enable_if_t<__ESIMD_DNS::get_num_args<Op>) == 1, simd<T, N> > sycl::_V1::ext::intel::esimd::atomic_update | ( | T * | p, |
| simd< Toffset, N > | offset, | ||
| simd< T, N > | src0, | ||
| simd_mask< N > | mask | ||
| ) |
LSC version of the single-argument atomic update.
Definition at line 3131 of file memory.hpp.
| __ESIMD_API std::enable_if_t<__ESIMD_DNS::get_num_args<Op>) == 0, simd<T, N> > sycl::_V1::ext::intel::esimd::atomic_update | ( | T * | p, |
| simd< Toffset, N > | offset, | ||
| simd_mask< N > | mask | ||
| ) |
LSC version of no argument variant of the atomic_update - accepts native::lsc::atomic_op instead of atomic_op as atomic operation template argument.
Definition at line 3104 of file memory.hpp.
| __ESIMD_API std::enable_if_t<__ESIMD_DNS::get_num_args<Op>) == 2, simd<T, N> > sycl::_V1::ext::intel::esimd::atomic_update | ( | T * | p, |
| simd_view< Toffset, RegionTy > | offsets, | ||
| simd< T, N > | src0, | ||
| simd< T, N > | src1, | ||
| simd_mask< N > | mask = 1 |
||
| ) |
Definition at line 3171 of file memory.hpp.
| __ESIMD_API __ESIMD_API std::enable_if_t<__ESIMD_DNS::get_num_args<Op>) == 1, simd<T, N> > sycl::_V1::ext::intel::esimd::atomic_update | ( | T * | p, |
| simd_view< Toffset, RegionTy > | offsets, | ||
| simd< T, N > | src0, | ||
| simd_mask< N > | mask = 1 |
||
| ) |
Definition at line 3141 of file memory.hpp.
| __ESIMD_API std::enable_if_t<__ESIMD_DNS::get_num_args<Op>) == 0, simd<T, N> > sycl::_V1::ext::intel::esimd::atomic_update | ( | T * | p, |
| simd_view< Toffset, RegionTy > | offsets, | ||
| simd_mask< N > | mask = 1 |
||
| ) |
Definition at line 3112 of file memory.hpp.
| __ESIMD_API std::enable_if_t<std::is_integral_v<Toffset> && __ESIMD_DNS::get_num_args<Op>) == 2, sycl::ext::intel::esimd::simd<T, N> > sycl::_V1::ext::intel::esimd::atomic_update | ( | T * | p, |
| Toffset | offset, | ||
| simd< T, N > | src0, | ||
| simd< T, N > | src1, | ||
| simd_mask< N > | mask = 1 |
||
| ) |
Definition at line 3181 of file memory.hpp.
| __ESIMD_API std::enable_if_t<std::is_integral_v<Toffset> && __ESIMD_DNS::get_num_args<Op>) == 1, simd<T, N> > sycl::_V1::ext::intel::esimd::atomic_update | ( | T * | p, |
| Toffset | offset, | ||
| simd< T, N > | src0, | ||
| simd_mask< N > | mask = 1 |
||
| ) |
Definition at line 3151 of file memory.hpp.
| __ESIMD_API std::enable_if_t<std::is_integral_v<Toffset> && __ESIMD_DNS::get_num_args<Op>) == 0, simd<T, N> > sycl::_V1::ext::intel::esimd::atomic_update | ( | T * | p, |
| Toffset | offset, | ||
| simd_mask< N > | mask = 1 |
||
| ) |
Definition at line 3122 of file memory.hpp.
| __ESIMD_API std::enable_if_t<std::is_integral<T>::value && !std::is_signed<T>::value && (sizeof(T) == 4), simd<T, N> > sycl::_V1::ext::intel::esimd::fbh | ( | simd< T, N > | src | ) |
Find the per element number of the first bit set in the source operand starting from the most significant bit (sign bit is counted).
| src0 | the source operand to count bits in. |
0xFFFFffff is returned for an element equal to 0. 
1.8.17