Base types

Defines vector, vector reference and matrix reference data types. More...

Collaboration diagram for Base types:

Concepts
concept	gpu::xetla::xetla_vector_ref
	Workaround for ESIMD vector(1D) ref type.
concept	gpu::xetla::xetla_matrix_ref
	Workaround for ESIMD matrix(2D) ref type.

Classes
struct	gpu::xetla::is_host_callable< T, typename >
struct	gpu::xetla::is_host_callable< T, std::enable_if_t< T::host_callable==true > >
struct	gpu::xetla::is_internal_type< T >
	Used to check if the type is xetla internal data type. More...
struct	gpu::xetla::is_floating_point< T >
	Used to check if the type is floating_point. More...
struct	gpu::xetla::is_integral< T >
	Used to check if the type is floating_point. More...
struct	gpu::xetla::native_type< T >
	Set the native data type of T. More...
struct	gpu::xetla::uint_type< T >
	Get the unit representation of type T. More...
struct	gpu::xetla::get_uint_type< Size >
	Get the unit representation based on Size. More...

Macros
#define	xetla_tdescriptor_ref   xetla_vector_ref<uint32_t, 16> __REF__
	Alias to xetla_vector<uint32_t, 16> reference.
#define	__REF__   auto
	Workaround for ESIMD reference usage.

Typedefs
using	gpu::xetla::bf16 = sycl::ext::oneapi::bfloat16
	xetla bf16 data type.
using	gpu::xetla::fp16 = sycl::half
	xetla fp16 data type.
using	gpu::xetla::tf32 = sycl::ext::intel::experimental::esimd::tfloat32
	xetla tf32 data type.
template<typename T >
using	gpu::xetla::native_type_t = typename native_type< T >::type
	Return the native data type of T.
template<typename T >
using	gpu::xetla::uint_type_t = typename uint_type< T >::type
	Return the uint representation of type T.
template<int Size>
using	gpu::xetla::get_uint_type_t = typename get_uint_type< Size >::type
	Return the uint representation based on Size.
template<typename Ty , uint32_t N>
using	gpu::xetla::xetla_vector = __ESIMD_NS::simd< native_type_t< Ty >, N >
	wrapper for xetla_vector.
using	gpu::xetla::xetla_tdescriptor = xetla_vector< uint32_t, 16 >
	Description of nd tensor descriptor for load and store.
template<uint32_t N>
using	gpu::xetla::xetla_mask = __ESIMD_NS::simd_mask< N >
	wrapper for xetla_mask.
template<uint32_t N>
using	gpu::xetla::xetla_mask_int = __ESIMD_NS::simd_mask< N >
	wrapper for xetla_mask_int.

Detailed Description

Defines vector, vector reference and matrix reference data types.

Macro Definition Documentation

__REF__

#define __REF__   auto

Workaround for ESIMD reference usage.

Alias to auto if go with ESIMD path.

See also

gpu::xetla::core::xetla_matrix_ref gpu::xetla::core::xetla_vector_ref

xetla_tdescriptor_ref

#define xetla_tdescriptor_ref   xetla_vector_ref<uint32_t, 16> __REF__

Alias to xetla_vector<uint32_t, 16> reference.

Typedef Documentation

bf16

using gpu::xetla::bf16 = typedef sycl::ext::oneapi::bfloat16

xetla bf16 data type.

The difference between bf16 and fp32 is:

fp32: 0_00000000_00000000000000000000000

bf16: 0_00000000_0000000

Note

The member function in bf16 class is only used in host side. For device side, we will automatically convert it to its native type.

See also

native_type_t

fp16

using gpu::xetla::fp16 = typedef sycl::half

xetla fp16 data type.

get_uint_type_t

template<int Size>

using gpu::xetla::get_uint_type_t = typedef typename get_uint_type<Size>::type

Return the uint representation based on Size.

native_type_t

template<typename T >

using gpu::xetla::native_type_t = typedef typename native_type<T>::type

Return the native data type of T.

tf32

using gpu::xetla::tf32 = typedef sycl::ext::intel::experimental::esimd::tfloat32

xetla tf32 data type.

The difference between tf32 and fp32 is:

fp32: 0_00000000_00000000000000000000000

tf32: 0_00000000_0000000000

Note

The member function in tf32 class is only used in host side. For device side, we will automatically convert it to its native type.

See also

native_type_t

uint_type_t

template<typename T >

using gpu::xetla::uint_type_t = typedef typename uint_type<T>::type

Return the uint representation of type T.

xetla_mask

template<uint32_t N>

using gpu::xetla::xetla_mask = typedef __ESIMD_NS::simd_mask<N>

wrapper for xetla_mask.

Alias to ESIMD __ESIMD_NS::simd_mask.

Template Parameters

N	data length in xetla_mask.

xetla_mask_int

template<uint32_t N>

using gpu::xetla::xetla_mask_int = typedef __ESIMD_NS::simd_mask<N>

wrapper for xetla_mask_int.

Alias to ESIMD __ESIMD_NS::simd_mask.

Template Parameters

N	data length in xetla_mask_int.

xetla_tdescriptor

using gpu::xetla::xetla_tdescriptor = typedef xetla_vector<uint32_t, 16>

Description of nd tensor descriptor for load and store.

Structure is defined in here.

xetla_vector

template<typename Ty , uint32_t N>

using gpu::xetla::xetla_vector = typedef __ESIMD_NS::simd<native_type_t<Ty>, N>

wrapper for xetla_vector.

Alias to ESIMD __ESIMD_NS::simd;

Template Parameters

Ty	data type in xetla_vector.
N	data length in xetla_vector.