DPC++ Runtime
Runtime libraries for oneAPI DPC++

Core APIs defining main vector data types and their interfaces. More...

Collaboration diagram for ESIMD core.:

Modules

 C++ binary operators overloads for ESIMD.
 Standard C++ binary operators overloads applicable to simd_obj_impl derivatives - simd , simd_mask , simd_view and their combinations.
 
 Alignment control.
 Alignment type tags and related APIs for use with ESIMD memory access operations.
 
 Main vector data types.
 ESIMD defines the following two main vector data types:
 

Classes

struct  sycl::_V1::ext::intel::esimd::saturation_on_tag
 Gen hardware supports applying saturation to results of certain operations. More...
 
struct  sycl::_V1::ext::intel::esimd::saturation_off_tag
 This type tag represents "saturation off" behavior. More...
 

Typedefs

using sycl::_V1::ext::intel::esimd::uchar = unsigned char
 
using sycl::_V1::ext::intel::esimd::ushort = unsigned short
 
using sycl::_V1::ext::intel::esimd::uint = unsigned int
 
using sycl::_V1::ext::intel::esimd::SurfaceIndex = unsigned int
 Surface index type. More...
 
using sycl::_V1::ext::intel::experimental::esimd::lsc_data_size = ("use sycl::ext::intel::esimd::memory_kind") lsc_memory_kind __ESIMD_DNS::lsc_data_size
 The scope that lsc_fence operation should apply to Supported platforms: DG2, PVC. More...
 
using sycl::_V1::ext::intel::experimental::esimd::cache_hint = sycl::ext::intel::esimd::cache_hint
 L1 or L2 cache hint kinds. More...
 

Enumerations

enum class  sycl::_V1::ext::intel::esimd::rgba_channel : uint8_t { sycl::_V1::ext::intel::esimd::R , sycl::_V1::ext::intel::esimd::G , sycl::_V1::ext::intel::esimd::B , sycl::_V1::ext::intel::esimd::A }
 Represents a pixel's channel. More...
 
enum class  sycl::_V1::ext::intel::esimd::raw_send_eot : uint8_t { sycl::_V1::ext::intel::esimd::not_eot = 0 , sycl::_V1::ext::intel::esimd::eot = 1 }
 Specify if end of thread should be set. More...
 
enum class  sycl::_V1::ext::intel::esimd::raw_send_sendc : uint8_t { sycl::_V1::ext::intel::esimd::not_sendc = 0 , sycl::_V1::ext::intel::esimd::sendc = 1 }
 Specify if sendc should be used. More...
 
enum class  sycl::_V1::ext::intel::esimd::rgba_channel_mask : uint8_t {
  sycl::_V1::ext::intel::esimd::R = detail::chR , sycl::_V1::ext::intel::esimd::G = detail::chG , sycl::_V1::ext::intel::esimd::GR = detail::chG | detail::chR , sycl::_V1::ext::intel::esimd::B = detail::chB ,
  sycl::_V1::ext::intel::esimd::BR = detail::chB | detail::chR , sycl::_V1::ext::intel::esimd::BG = detail::chB | detail::chG , sycl::_V1::ext::intel::esimd::BGR = detail::chB | detail::chG | detail::chR , sycl::_V1::ext::intel::esimd::A = detail::chA ,
  sycl::_V1::ext::intel::esimd::AR = detail::chA | detail::chR , sycl::_V1::ext::intel::esimd::AG = detail::chA | detail::chG , sycl::_V1::ext::intel::esimd::AGR = detail::chA | detail::chG | detail::chR , sycl::_V1::ext::intel::esimd::AB = detail::chA | detail::chB ,
  sycl::_V1::ext::intel::esimd::ABR = detail::chA | detail::chB | detail::chR , sycl::_V1::ext::intel::esimd::ABG = detail::chA | detail::chB | detail::chG , sycl::_V1::ext::intel::esimd::ABGR = detail::chA | detail::chB | detail::chG | detail::chR
}
 Represents a pixel's channel mask - all possible combinations of enabled channels. More...
 
enum class  sycl::_V1::ext::intel::esimd::atomic_op : uint8_t {
  sycl::_V1::ext::intel::esimd::add = 0x0 , sycl::_V1::ext::intel::esimd::sub = 0x1 , sycl::_V1::ext::intel::esimd::inc = 0x2 , sycl::_V1::ext::intel::esimd::dec = 0x3 ,
  sycl::_V1::ext::intel::esimd::umin = 0x4 , sycl::_V1::ext::intel::esimd::umax = 0x5 , sycl::_V1::ext::intel::esimd::xchg = 0x6 , sycl::_V1::ext::intel::esimd::cmpxchg = 0x7 ,
  sycl::_V1::ext::intel::esimd::bit_and = 0x8 , sycl::_V1::ext::intel::esimd::bit_or = 0x9 , sycl::_V1::ext::intel::esimd::bit_xor = 0xa , sycl::_V1::ext::intel::esimd::smin = 0xb ,
  sycl::_V1::ext::intel::esimd::smax = 0xc , sycl::_V1::ext::intel::esimd::fmax = 0x10 , sycl::_V1::ext::intel::esimd::fmin = 0x11 , sycl::_V1::ext::intel::esimd::fcmpxchg = 0x12 ,
  sycl::_V1::ext::intel::esimd::fcmpwr = fcmpxchg , sycl::_V1::ext::intel::esimd::fadd = 0x13 , sycl::_V1::ext::intel::esimd::fsub = 0x14 , sycl::_V1::ext::intel::esimd::load = 0x15 ,
  sycl::_V1::ext::intel::esimd::store = 0x16 , sycl::_V1::ext::intel::esimd::predec = 0xff
}
 Represents an atomic operation. More...
 
enum class  sycl::_V1::ext::intel::experimental::esimd::split_barrier_action : uint8_t { sycl::_V1::ext::intel::experimental::esimd::wait = 0 , sycl::_V1::ext::intel::experimental::esimd::signal = 1 }
 Represents a split barrier action. More...
 

Functions

constexpr int sycl::_V1::ext::intel::esimd::is_channel_enabled (rgba_channel_mask M, rgba_channel Ch)
 
constexpr int sycl::_V1::ext::intel::esimd::get_num_channels_enabled (rgba_channel_mask M)
 

Variables

static constexpr saturation_off_tag sycl::_V1::ext::intel::esimd::saturation_off {}
 Type tag object representing "saturation off" behavior. More...
 
static constexpr saturation_on_tag sycl::_V1::ext::intel::esimd::saturation_on {}
 Type tag object representing "saturation on" behavior. More...
 

Detailed Description

Core APIs defining main vector data types and their interfaces.

Typedef Documentation

◆ cache_hint

L1 or L2 cache hint kinds.

Definition at line 108 of file common.hpp.

◆ lsc_data_size

using sycl::_V1::ext::intel::experimental::esimd::lsc_data_size = typedef ("use sycl::ext::intel::esimd::memory_kind") lsc_memory_kind __ESIMD_DNS::lsc_data_size
strong

The scope that lsc_fence operation should apply to Supported platforms: DG2, PVC.

The lsc_fence operation to apply to caches Supported platforms: DG2, PVC The specific LSC shared function to fence with lsc_fence Supported platforms: DG2, PVC

Definition at line 64 of file common.hpp.

◆ SurfaceIndex

using sycl::_V1::ext::intel::esimd::SurfaceIndex = typedef unsigned int

Surface index type.

Surface is an internal representation of a memory block addressable by GPU in "stateful" memory model, and each surface is identified by its "binding table index" - surface index.

Definition at line 64 of file common.hpp.

◆ uchar

using sycl::_V1::ext::intel::esimd::uchar = typedef unsigned char

Definition at line 41 of file common.hpp.

◆ uint

using sycl::_V1::ext::intel::esimd::uint = typedef unsigned int

Definition at line 43 of file common.hpp.

◆ ushort

using sycl::_V1::ext::intel::esimd::ushort = typedef unsigned short

Definition at line 42 of file common.hpp.

Enumeration Type Documentation

◆ atomic_op

Represents an atomic operation.

Operations always return the old value(s) of the target memory location(s) as it was before the operation was applied. Each operation is annotated with a pseudocode illustrating its semantics, addr is a memory address (one of the many, as the atomic operation is vector) the operation is applied at, src0 is its first argumnet, src1 - second. Using the floating point atomic operations adds the requirement to running the code with it on target devices with LSC features (ACM, PVC, etc).

Enumerator
add 

Addition: *addr = *addr + src0.

sub 

Subtraction: *addr = *addr - src0.

inc 

Increment: *addr = *addr + 1.

dec 

Decrement: *addr = *addr - 1.

umin 

Minimum: *addr = min(*addr, src0).

umax 

Maximum: *addr = max(*addr, src0).

xchg 

Exchange. *addr == src0;

cmpxchg 

Compare and exchange. if (*addr == src0) *sddr = src1;

bit_and 

Bit and: *addr = *addr & src0.

bit_or 

Bit or: *addr = *addr | src0.

bit_xor 

Bit xor: *addr = *addr | src0.

smin 

Minimum (signed integer): *addr = min(*addr, src0).

smax 

Maximum (signed integer): *addr = max(*addr, src0).

fmax 

ACM/PVC: Minimum (floating point): *addr = min(*addr, src0).

fmin 

ACM/PVC: Maximum (floating point): *addr = max(*addr, src0).

fcmpxchg 

ACM/PVC: Compare and exchange (floating point).

if (*addr == src0) *addr = src1;

fcmpwr 
fadd 

ACM/PVC: Addition (floating point): *addr = *addr + src0.

fsub 

ACM/PVC: Subtraction (floating point): *addr = *addr - src0.

load 
store 
predec 

Decrement: *addr = *addr - 1.

The only operation which returns new value of the destination rather than old.

Definition at line 160 of file common.hpp.

◆ raw_send_eot

Specify if end of thread should be set.

Enumerator
not_eot 
eot 

Definition at line 67 of file common.hpp.

◆ raw_send_sendc

Specify if sendc should be used.

Enumerator
not_sendc 
sendc 

Definition at line 73 of file common.hpp.

◆ rgba_channel

Represents a pixel's channel.

Enumerator

Definition at line 59 of file common.hpp.

◆ rgba_channel_mask

Represents a pixel's channel mask - all possible combinations of enabled channels.

Enumerator
GR 
BR 
BG 
BGR 
AR 
AG 
AGR 
AB 
ABR 
ABG 
ABGR 

Definition at line 122 of file common.hpp.

◆ split_barrier_action

Represents a split barrier action.

Enumerator
wait 
signal 

Definition at line 111 of file common.hpp.

Function Documentation

◆ get_num_channels_enabled()

constexpr int sycl::_V1::ext::intel::esimd::get_num_channels_enabled ( rgba_channel_mask  M)
constexpr

◆ is_channel_enabled()

constexpr int sycl::_V1::ext::intel::esimd::is_channel_enabled ( rgba_channel_mask  M,
rgba_channel  Ch 
)
constexpr

Definition at line 140 of file common.hpp.

Referenced by sycl::_V1::ext::intel::esimd::get_num_channels_enabled().

Variable Documentation

◆ saturation_off

constexpr saturation_off_tag sycl::_V1::ext::intel::esimd::saturation_off {}
inlinestaticconstexpr

Type tag object representing "saturation off" behavior.

Definition at line 53 of file common.hpp.

◆ saturation_on

constexpr saturation_on_tag sycl::_V1::ext::intel::esimd::saturation_on {}
inlinestaticconstexpr

Type tag object representing "saturation on" behavior.

Definition at line 56 of file common.hpp.