DPC++ Runtime
Runtime libraries for oneAPI DPC++
sycl::_V1::ext::oneapi::experimental::matrix Namespace Reference

Namespaces

 precision
 

Classes

struct  joint_matrix
 
struct  spv_matrix_layout_traits
 
struct  spv_scope_traits
 
struct  spv_scope_traits< sycl::group< D > >
 
struct  spv_scope_traits< sycl::sub_group >
 
struct  tpu_params
 
struct  tpu_params< tpu::amx, Ta, Tb, Tc, 0, 0, 0, typename std::enable_if<(!std::is_same_v< Ta, void > &&!std::is_same_v< Tb, void > &&!std::is_same_v< Tc, void >)>::type >
 
struct  tpu_params< tpu::amx, Ta, Tb, Tc, M, N, K, typename std::enable_if<(!std::is_same_v< Ta, void > &&!std::is_same_v< Tb, void > &&!std::is_same_v< Tc, void > &&M !=0 &&N !=0 &&K !=0)>::type >
 
struct  tpu_params< tpu::amx, Ta, Tb, Tc, sM, sN, sK, typename std::enable_if<(!std::is_same_v< Ta, void > &&!std::is_same_v< Tb, void > &&!std::is_same_v< Tc, void > &&sM !=0 &&sN !=0 &&sK !=0)>::type >
 
struct  tpu_params< tpu::amx, void, void, void, M, N, K >
 
struct  tpu_params< tpu::amx, void, void, void, sM, sN, sK >
 
struct  tpu_params< tpu::dpas, Ta, Tb, Tc, 0, 0, 0, typename std::enable_if<(!std::is_same_v< Ta, void > &&!std::is_same_v< Tb, void > &&!std::is_same_v< Tc, void >)>::type >
 
struct  tpu_params< tpu::dpas, Ta, Tb, Tc, M, N, K, typename std::enable_if<((!std::is_same_v< Ta, void > &&M !=0))>::type >
 
struct  tpu_params< tpu::dpas, void, void, void, M, N, K >
 
struct  tpu_params< tpu::xmx16, Ta, Tb, Tc, 0, 0, 0, typename std::enable_if<(!std::is_same_v< Ta, void > &&!std::is_same_v< Tb, void > &&!std::is_same_v< Tc, void >)>::type >
 
struct  tpu_params< tpu::xmx16, Ta, Tb, Tc, sM, sN, sK, typename std::enable_if<((!std::is_same_v< Ta, void > &&sM !=0))>::type >
 
struct  tpu_params< tpu::xmx16, void, void, void, sM, sN, sK >
 
struct  tpu_params< tpu::xmx8, Ta, Tb, Tc, 0, 0, 0, typename std::enable_if<(!std::is_same_v< Ta, void > &&!std::is_same_v< Tb, void > &&!std::is_same_v< Tc, void >)>::type >
 
struct  tpu_params< tpu::xmx8, Ta, Tb, Tc, sM, sN, sK, typename std::enable_if<((!std::is_same_v< Ta, void > &&sM !=0))>::type >
 
struct  tpu_params< tpu::xmx8, void, void, void, sM, sN, sK >
 
class  wi_data
 
class  wi_element
 
class  wi_element< sycl::ext::oneapi::bfloat16, NumRows, NumCols, Layout, Group >
 
class  wi_element< sycl::ext::oneapi::bfloat16, NumRows, NumCols, Use, Layout, Group >
 
class  wi_element< uint16_t, NumRows, NumCols, Layout, Group >
 

Enumerations

enum  matrix_layout {
  matrix_layout::row_major, matrix_layout::col_major, matrix_layout::packed_a, matrix_layout::packed_b,
  matrix_layout::row_major, matrix_layout::col_major, matrix_layout::packed_a, matrix_layout::packed_b
}
 
enum  matrix_use { matrix_use::a, matrix_use::b, matrix_use::accumulator }
 
enum  matrix_layout {
  matrix_layout::row_major, matrix_layout::col_major, matrix_layout::packed_a, matrix_layout::packed_b,
  matrix_layout::row_major, matrix_layout::col_major, matrix_layout::packed_a, matrix_layout::packed_b
}
 
enum  use { use::a, use::b, use::accumulator }
 
enum  layout { layout::row_major = 0, layout::col_major = 1, layout::dynamic = 3 }
 
enum  tpu {
  tpu::xmx8, tpu::xmx16, tpu::amx, tpu::dpas,
  tpu::amx
}
 
enum  matrix_type {
  matrix_type::bf8, matrix_type::bf16, matrix_type::fp16, matrix_type::tf32,
  matrix_type::fp32, matrix_type::fp64, matrix_type::sint2, matrix_type::sint4,
  matrix_type::sint8, matrix_type::sint16, matrix_type::sint32, matrix_type::sint64,
  matrix_type::uint2, matrix_type::uint4, matrix_type::uint8, matrix_type::uint16,
  matrix_type::uint32, matrix_type::uint64, matrix_type::bf8, matrix_type::bf16,
  matrix_type::fp16, matrix_type::fp19, matrix_type::fp32, matrix_type::fp64,
  matrix_type::sint2, matrix_type::sint4, matrix_type::sint8, matrix_type::sint16,
  matrix_type::sint32, matrix_type::sint64, matrix_type::uint2, matrix_type::uint4,
  matrix_type::uint8, matrix_type::uint16, matrix_type::uint32, matrix_type::uint64
}
 
enum  scope_t { scope_t::sub_group, scope_t::work_group, scope_t::sub_group, scope_t::work_group }
 
enum  tpu {
  tpu::xmx8, tpu::xmx16, tpu::amx, tpu::dpas,
  tpu::amx
}
 
enum  matrix_type {
  matrix_type::bf8, matrix_type::bf16, matrix_type::fp16, matrix_type::tf32,
  matrix_type::fp32, matrix_type::fp64, matrix_type::sint2, matrix_type::sint4,
  matrix_type::sint8, matrix_type::sint16, matrix_type::sint32, matrix_type::sint64,
  matrix_type::uint2, matrix_type::uint4, matrix_type::uint8, matrix_type::uint16,
  matrix_type::uint32, matrix_type::uint64, matrix_type::bf8, matrix_type::bf16,
  matrix_type::fp16, matrix_type::fp19, matrix_type::fp32, matrix_type::fp64,
  matrix_type::sint2, matrix_type::sint4, matrix_type::sint8, matrix_type::sint16,
  matrix_type::sint32, matrix_type::sint64, matrix_type::uint2, matrix_type::uint4,
  matrix_type::uint8, matrix_type::uint16, matrix_type::uint32, matrix_type::uint64
}
 
enum  scope_t { scope_t::sub_group, scope_t::work_group, scope_t::sub_group, scope_t::work_group }
 

Functions

 SPV_MATRIX_LAYOUT_TRAITS (sycl::ext::intel::experimental::matrix::layout::packed, __spv::MatrixLayout::Packed) template< use Use > struct spv_matrix_use_traits
 
template<typename Group , typename T , size_t NumRows, size_t NumCols, matrix_layout Layout = matrix_layout::row_major, access::address_space Space, access::decorated IsDecorated>
__SYCL_ALWAYS_INLINE void joint_matrix_load (Group sg, joint_matrix< T, NumRows, NumCols, Layout, Group > &res, multi_ptr< T, Space, IsDecorated > src, size_t stride, matrix_layout MemL)
 
template<typename Group , typename T , size_t NumRows, size_t NumCols, matrix_layout MatL = matrix_layout::row_major, access::address_space Space, access::decorated IsDecorated>
__SYCL_ALWAYS_INLINE void joint_matrix_store (Group sg, joint_matrix< T, NumRows, NumCols, MatL, Group > &src, multi_ptr< T, Space, IsDecorated > res, size_t stride, matrix_layout MemL)
 
template<typename Group , typename T1 , typename T2 , typename T3 , size_t M, size_t K, size_t N, matrix_layout LayoutA, matrix_layout LayoutB, matrix_layout LayoutC>
__SYCL_ALWAYS_INLINE joint_matrix< T3, M, N, LayoutC, Group > joint_matrix_mad (Group sg, joint_matrix< T1, M, K, LayoutA, Group > &mA, joint_matrix< T2, K, N, LayoutB, Group > &mB, joint_matrix< T3, M, N, LayoutC, Group > &mC)
 
template<typename Group , typename T , size_t NumRows, size_t NumCols, matrix_layout Layout, typename T2 >
__SYCL_ALWAYS_INLINE void joint_matrix_fill (Group sg, joint_matrix< T, NumRows, NumCols, Layout, Group > &res, const T2 v)
 
template<typename Group , typename T , matrix_use Use, size_t NumRows, size_t NumCols, matrix_layout Layout, typename T2 >
__SYCL_ALWAYS_INLINE void joint_matrix_fill (Group sg, joint_matrix< T, Use, NumRows, NumCols, Layout, Group > &res, const T2 v)
 
template<typename Group , typename S , typename T , matrix_use Use, size_t NumRows, size_t NumCols, matrix_layout Layout, access::address_space Space, access::decorated IsDecorated, std::enable_if_t< std::is_same< S, std::remove_const_t< T >>::value||(std::is_same< S, precision::tf32 >::value &&std::is_same< std::remove_const_t< T >, float >::value), bool > = true>
void joint_matrix_load (Group sg, joint_matrix< S, Use, NumRows, NumCols, Layout, Group > &res, multi_ptr< T, Space, IsDecorated > src, size_t stride)
 
template<typename Group , typename T , size_t NumRows, size_t NumCols, matrix_layout Layout, access::address_space Space, access::decorated IsDecorated>
void joint_matrix_store (Group sg, joint_matrix< T, matrix_use::accumulator, NumRows, NumCols, Layout, Group > &src, multi_ptr< T, Space, IsDecorated > dst, size_t stride)
 
template<typename Group , typename T1 , typename T2 , std::size_t M, std::size_t K, std::size_t N, matrix_layout LayoutA, matrix_layout LayoutB, matrix_layout LayoutC>
joint_matrix< T2, matrix_use::accumulator, M, N, LayoutC, Group > joint_matrix_mad (Group sg, joint_matrix< T1, matrix_use::a, M, K, LayoutA, Group > A, joint_matrix< T1, matrix_use::b, K, N, LayoutB, Group > B, joint_matrix< T2, matrix_use::accumulator, M, N, LayoutC, Group > C)
 
__SYCL_ALWAYS_INLINE float round_to_tf32 (float a)
 
template<typename Group , typename T , use Use, size_t Rows, size_t Cols, layout Layout>
decltype(auto) __SYCL_ALWAYS_INLINE get_wi_data (Group sg, joint_matrix< Group, T, Use, Rows, Cols, Layout > &jm)
 
template<typename Group , typename T , size_t NumRows, size_t NumCols, use Use, layout Layout, typename T2 >
__SYCL_ALWAYS_INLINE void joint_matrix_fill (Group sg, joint_matrix< Group, T, Use, NumRows, NumCols, Layout > &res, const T2 &v)
 
template<typename Group , typename S , typename T , size_t NumRows, size_t NumCols, access::address_space Space, access::decorated IsDecorated, std::enable_if_t< std::is_same< S, std::remove_const_t< T >>::value, bool > = true>
__SYCL_ALWAYS_INLINE void joint_matrix_load (Group sg, joint_matrix< Group, S, use::accumulator, NumRows, NumCols, sycl::ext::oneapi::experimental::matrix::layout::dynamic > &res, multi_ptr< T, Space, IsDecorated > src, size_t stride, sycl::ext::oneapi::experimental::matrix::layout Layout)
 
template<typename Group , typename S , typename T , use Use, size_t NumRows, size_t NumCols, matrix::layout Layout, access::address_space Space, access::decorated IsDecorated, std::enable_if_t< std::is_same< S, std::remove_const_t< T >>::value||(std::is_same< S, precision::tf32 >::value &&std::is_same< std::remove_const_t< T >, float >::value), bool > = true>
__SYCL_ALWAYS_INLINE void joint_matrix_load (Group sg, joint_matrix< Group, S, Use, NumRows, NumCols, Layout > &res, multi_ptr< T, Space, IsDecorated > src, size_t stride)
 
template<typename Group , typename T , size_t NumRows, size_t NumCols, access::address_space Space, access::decorated IsDecorated>
__SYCL_ALWAYS_INLINE void joint_matrix_store (Group sg, joint_matrix< Group, T, use::accumulator, NumRows, NumCols, sycl::ext::oneapi::experimental::matrix::layout::dynamic > &src, multi_ptr< T, Space, IsDecorated > dst, size_t stride, sycl::ext::oneapi::experimental::matrix::layout Layout)
 
template<typename Group , typename Ta , typename Tb , typename Tc , std::size_t M, std::size_t K, std::size_t N, layout LayoutA, layout LayoutB>
__SYCL_ALWAYS_INLINE joint_matrix< Group, Tc, use::accumulator, M, N, sycl::ext::oneapi::experimental::matrix::layout::dynamic > joint_matrix_mad (Group sg, joint_matrix< Group, Ta, use::a, M, K, LayoutA > &A, joint_matrix< Group, Tb, use::b, K, N, LayoutB > &B, joint_matrix< Group, Tc, use::accumulator, M, N, sycl::ext::oneapi::experimental::matrix::layout::dynamic > &C)
 
__SYCL_ALWAYS_INLINE float round_to_tf32 (float &a)
 
template<typename Ta , typename Tb , typename Tc >
constexpr bool is_combination_valid_amx (int sM, int sN, int sK)
 
template<typename Ta , typename Tb , typename Tc >
constexpr bool are_types_valid_amx ()
 
template<typename Ta , typename Tb , typename Tc >
constexpr bool is_combination_valid_xmx8 (int sM, int sN, int sK)
 
template<typename Ta , typename Tb , typename Tc >
constexpr bool are_types_valid_xmx8 ()
 
template<typename Ta , typename Tb , typename Tc >
constexpr bool is_combination_valid_xmx16 (int sM, int sN, int sK)
 
template<typename Ta , typename Tb , typename Tc >
constexpr bool are_types_valid_xmx16 ()
 
template<typename Ta , typename Tb , typename Tc >
constexpr bool is_combination_valid_dpas (int M, int N, int K)
 
template<typename Ta , typename Tb , typename Tc >
constexpr bool are_types_valid_dpas ()
 

Enumeration Type Documentation

◆ layout

Enumerator
row_major 
col_major 
dynamic 

Definition at line 19 of file matrix-unified-utils.hpp.

◆ matrix_layout [1/2]

Enumerator
row_major 
col_major 
packed_a 
packed_b 
row_major 
col_major 
packed_a 
packed_b 

Definition at line 20 of file matrix-jit.hpp.

◆ matrix_layout [2/2]

Enumerator
row_major 
col_major 
packed_a 
packed_b 
row_major 
col_major 
packed_a 
packed_b 

Definition at line 20 of file matrix-tensorcores-legacy.hpp.

◆ matrix_type [1/2]

Enumerator
bf8 
bf16 
fp16 
tf32 
fp32 
fp64 
sint2 
sint4 
sint8 
sint16 
sint32 
sint64 
uint2 
uint4 
uint8 
uint16 
uint32 
uint64 
bf8 
bf16 
fp16 
fp19 
fp32 
fp64 
sint2 
sint4 
sint8 
sint16 
sint32 
sint64 
uint2 
uint4 
uint8 
uint16 
uint32 
uint64 

Definition at line 34 of file static-query.hpp.

◆ matrix_type [2/2]

Enumerator
bf8 
bf16 
fp16 
tf32 
fp32 
fp64 
sint2 
sint4 
sint8 
sint16 
sint32 
sint64 
uint2 
uint4 
uint8 
uint16 
uint32 
uint64 
bf8 
bf16 
fp16 
fp19 
fp32 
fp64 
sint2 
sint4 
sint8 
sint16 
sint32 
sint64 
uint2 
uint4 
uint8 
uint16 
uint32 
uint64 

Definition at line 34 of file static-query-use.hpp.

◆ matrix_use

Enumerator
accumulator 

Definition at line 18 of file matrix-tensorcores-legacy.hpp.

◆ scope_t [1/2]

Enumerator
sub_group 
work_group 
sub_group 
work_group 

Definition at line 55 of file static-query-use.hpp.

◆ scope_t [2/2]

Enumerator
sub_group 
work_group 
sub_group 
work_group 

Definition at line 55 of file static-query.hpp.

◆ tpu [1/2]

Enumerator
xmx8 
xmx16 
amx 
dpas 
amx 

Definition at line 29 of file static-query-use.hpp.

◆ tpu [2/2]

Enumerator
xmx8 
xmx16 
amx 
dpas 
amx 

Definition at line 30 of file static-query.hpp.

◆ use

Enumerator
accumulator 

Definition at line 17 of file matrix-unified-utils.hpp.

Function Documentation

◆ are_types_valid_amx()

template<typename Ta , typename Tb , typename Tc >
constexpr bool sycl::_V1::ext::oneapi::experimental::matrix::are_types_valid_amx ( )
constexpr

Definition at line 82 of file static-query-use.hpp.

◆ are_types_valid_dpas()

template<typename Ta , typename Tb , typename Tc >
constexpr bool sycl::_V1::ext::oneapi::experimental::matrix::are_types_valid_dpas ( )
constexpr

Definition at line 250 of file static-query.hpp.

◆ are_types_valid_xmx16()

template<typename Ta , typename Tb , typename Tc >
constexpr bool sycl::_V1::ext::oneapi::experimental::matrix::are_types_valid_xmx16 ( )
constexpr

Definition at line 434 of file static-query-use.hpp.

◆ are_types_valid_xmx8()

template<typename Ta , typename Tb , typename Tc >
constexpr bool sycl::_V1::ext::oneapi::experimental::matrix::are_types_valid_xmx8 ( )
constexpr

Definition at line 243 of file static-query-use.hpp.

◆ get_wi_data()

template<typename Group , typename T , use Use, size_t Rows, size_t Cols, layout Layout>
decltype(auto) __SYCL_ALWAYS_INLINE sycl::_V1::ext::oneapi::experimental::matrix::get_wi_data ( Group  sg,
joint_matrix< Group, T, Use, Rows, Cols, Layout > &  jm 
)
inline

Definition at line 98 of file matrix-unified.hpp.

◆ is_combination_valid_amx()

template<typename Ta , typename Tb , typename Tc >
constexpr bool sycl::_V1::ext::oneapi::experimental::matrix::is_combination_valid_amx ( int  sM,
int  sN,
int  sK 
)
constexpr

Definition at line 62 of file static-query-use.hpp.

◆ is_combination_valid_dpas()

template<typename Ta , typename Tb , typename Tc >
constexpr bool sycl::_V1::ext::oneapi::experimental::matrix::is_combination_valid_dpas ( int  M,
int  N,
int  K 
)
constexpr

Definition at line 225 of file static-query.hpp.

◆ is_combination_valid_xmx16()

template<typename Ta , typename Tb , typename Tc >
constexpr bool sycl::_V1::ext::oneapi::experimental::matrix::is_combination_valid_xmx16 ( int  sM,
int  sN,
int  sK 
)
constexpr

Definition at line 409 of file static-query-use.hpp.

◆ is_combination_valid_xmx8()

template<typename Ta , typename Tb , typename Tc >
constexpr bool sycl::_V1::ext::oneapi::experimental::matrix::is_combination_valid_xmx8 ( int  sM,
int  sN,
int  sK 
)
constexpr

Definition at line 218 of file static-query-use.hpp.

◆ joint_matrix_fill() [1/3]

template<typename Group , typename T , size_t NumRows, size_t NumCols, use Use, layout Layout, typename T2 >
__SYCL_ALWAYS_INLINE void sycl::_V1::ext::oneapi::experimental::matrix::joint_matrix_fill ( Group  sg,
joint_matrix< Group, T, Use, NumRows, NumCols, Layout > &  res,
const T2 &  v 
)
inline

Definition at line 118 of file matrix-unified.hpp.

◆ joint_matrix_fill() [2/3]

template<typename Group , typename T , size_t NumRows, size_t NumCols, matrix_layout Layout, typename T2 >
__SYCL_ALWAYS_INLINE void sycl::_V1::ext::oneapi::experimental::matrix::joint_matrix_fill ( Group  sg,
joint_matrix< T, NumRows, NumCols, Layout, Group > &  res,
const T2  v 
)
inline

◆ joint_matrix_fill() [3/3]

template<typename Group , typename T , matrix_use Use, size_t NumRows, size_t NumCols, matrix_layout Layout, typename T2 >
__SYCL_ALWAYS_INLINE void sycl::_V1::ext::oneapi::experimental::matrix::joint_matrix_fill ( Group  sg,
joint_matrix< T, Use, NumRows, NumCols, Layout, Group > &  res,
const T2  v 
)
inline

Definition at line 160 of file matrix-tensorcores-legacy.hpp.

◆ joint_matrix_load() [1/4]

template<typename Group , typename S , typename T , use Use, size_t NumRows, size_t NumCols, matrix::layout Layout, access::address_space Space, access::decorated IsDecorated, std::enable_if_t< std::is_same< S, std::remove_const_t< T >>::value||(std::is_same< S, precision::tf32 >::value &&std::is_same< std::remove_const_t< T >, float >::value), bool > = true>
__SYCL_ALWAYS_INLINE void sycl::_V1::ext::oneapi::experimental::matrix::joint_matrix_load ( Group  sg,
joint_matrix< Group, S, Use, NumRows, NumCols, Layout > &  res,
multi_ptr< T, Space, IsDecorated >  src,
size_t  stride 
)
inline

◆ joint_matrix_load() [2/4]

template<typename Group , typename S , typename T , size_t NumRows, size_t NumCols, access::address_space Space, access::decorated IsDecorated, std::enable_if_t< std::is_same< S, std::remove_const_t< T >>::value, bool > = true>
__SYCL_ALWAYS_INLINE void sycl::_V1::ext::oneapi::experimental::matrix::joint_matrix_load ( Group  sg,
joint_matrix< Group, S, use::accumulator, NumRows, NumCols, sycl::ext::oneapi::experimental::matrix::layout::dynamic > &  res,
multi_ptr< T, Space, IsDecorated >  src,
size_t  stride,
sycl::ext::oneapi::experimental::matrix::layout  Layout 
)
inline

◆ joint_matrix_load() [3/4]

template<typename Group , typename S , typename T , matrix_use Use, size_t NumRows, size_t NumCols, matrix_layout Layout, access::address_space Space, access::decorated IsDecorated, std::enable_if_t< std::is_same< S, std::remove_const_t< T >>::value||(std::is_same< S, precision::tf32 >::value &&std::is_same< std::remove_const_t< T >, float >::value), bool > = true>
void sycl::_V1::ext::oneapi::experimental::matrix::joint_matrix_load ( Group  sg,
joint_matrix< S, Use, NumRows, NumCols, Layout, Group > &  res,
multi_ptr< T, Space, IsDecorated >  src,
size_t  stride 
)

Definition at line 694 of file matrix-tensorcores-legacy.hpp.

References sycl::_V1::Space.

◆ joint_matrix_load() [4/4]

template<typename Group , typename T , size_t NumRows, size_t NumCols, matrix_layout Layout = matrix_layout::row_major, access::address_space Space, access::decorated IsDecorated>
__SYCL_ALWAYS_INLINE void sycl::_V1::ext::oneapi::experimental::matrix::joint_matrix_load ( Group  sg,
joint_matrix< T, NumRows, NumCols, Layout, Group > &  res,
multi_ptr< T, Space, IsDecorated >  src,
size_t  stride,
matrix_layout  MemL 
)
inline

◆ joint_matrix_mad() [1/3]

template<typename Group , typename Ta , typename Tb , typename Tc , std::size_t M, std::size_t K, std::size_t N, layout LayoutA, layout LayoutB>
__SYCL_ALWAYS_INLINE joint_matrix<Group, Tc, use::accumulator, M, N, sycl::ext::oneapi::experimental::matrix::layout::dynamic> sycl::_V1::ext::oneapi::experimental::matrix::joint_matrix_mad ( Group  sg,
joint_matrix< Group, Ta, use::a, M, K, LayoutA > &  A,
joint_matrix< Group, Tb, use::b, K, N, LayoutB > &  B,
joint_matrix< Group, Tc, use::accumulator, M, N, sycl::ext::oneapi::experimental::matrix::layout::dynamic > &  C 
)
inline

Definition at line 291 of file matrix-unified.hpp.

◆ joint_matrix_mad() [2/3]

template<typename Group , typename T1 , typename T2 , typename T3 , size_t M, size_t K, size_t N, matrix_layout LayoutA, matrix_layout LayoutB, matrix_layout LayoutC>
__SYCL_ALWAYS_INLINE joint_matrix<T3, M, N, LayoutC, Group> sycl::_V1::ext::oneapi::experimental::matrix::joint_matrix_mad ( Group  sg,
joint_matrix< T1, M, K, LayoutA, Group > &  mA,
joint_matrix< T2, K, N, LayoutB, Group > &  mB,
joint_matrix< T3, M, N, LayoutC, Group > &  mC 
)
inline

◆ joint_matrix_mad() [3/3]

template<typename Group , typename T1 , typename T2 , std::size_t M, std::size_t K, std::size_t N, matrix_layout LayoutA, matrix_layout LayoutB, matrix_layout LayoutC>
joint_matrix<T2, matrix_use::accumulator, M, N, LayoutC, Group> sycl::_V1::ext::oneapi::experimental::matrix::joint_matrix_mad ( Group  sg,
joint_matrix< T1, matrix_use::a, M, K, LayoutA, Group >  A,
joint_matrix< T1, matrix_use::b, K, N, LayoutB, Group >  B,
joint_matrix< T2, matrix_use::accumulator, M, N, LayoutC, Group >  C 
)

Definition at line 740 of file matrix-tensorcores-legacy.hpp.

◆ joint_matrix_store() [1/3]

template<typename Group , typename T , size_t NumRows, size_t NumCols, access::address_space Space, access::decorated IsDecorated>
__SYCL_ALWAYS_INLINE void sycl::_V1::ext::oneapi::experimental::matrix::joint_matrix_store ( Group  sg,
joint_matrix< Group, T, use::accumulator, NumRows, NumCols, sycl::ext::oneapi::experimental::matrix::layout::dynamic > &  src,
multi_ptr< T, Space, IsDecorated >  dst,
size_t  stride,
sycl::ext::oneapi::experimental::matrix::layout  Layout 
)
inline

◆ joint_matrix_store() [2/3]

template<typename Group , typename T , size_t NumRows, size_t NumCols, matrix_layout Layout, access::address_space Space, access::decorated IsDecorated>
void sycl::_V1::ext::oneapi::experimental::matrix::joint_matrix_store ( Group  sg,
joint_matrix< T, matrix_use::accumulator, NumRows, NumCols, Layout, Group > &  src,
multi_ptr< T, Space, IsDecorated >  dst,
size_t  stride 
)

Definition at line 716 of file matrix-tensorcores-legacy.hpp.

References sycl::_V1::Space.

◆ joint_matrix_store() [3/3]

template<typename Group , typename T , size_t NumRows, size_t NumCols, matrix_layout MatL = matrix_layout::row_major, access::address_space Space, access::decorated IsDecorated>
__SYCL_ALWAYS_INLINE void sycl::_V1::ext::oneapi::experimental::matrix::joint_matrix_store ( Group  sg,
joint_matrix< T, NumRows, NumCols, MatL, Group > &  src,
multi_ptr< T, Space, IsDecorated >  res,
size_t  stride,
matrix_layout  MemL 
)
inline

◆ round_to_tf32() [1/2]

__SYCL_ALWAYS_INLINE float sycl::_V1::ext::oneapi::experimental::matrix::round_to_tf32 ( float &  a)
inline

Definition at line 340 of file matrix-unified.hpp.

References a, and sycl::_V1::detail::memcpy().

◆ round_to_tf32() [2/2]

__SYCL_ALWAYS_INLINE float sycl::_V1::ext::oneapi::experimental::matrix::round_to_tf32 ( float  a)
inline

Definition at line 762 of file matrix-tensorcores-legacy.hpp.

References a.

◆ SPV_MATRIX_LAYOUT_TRAITS()

sycl::_V1::ext::oneapi::experimental::matrix::SPV_MATRIX_LAYOUT_TRAITS ( sycl::ext::intel::experimental::matrix::layout::packed  ,
__spv::MatrixLayout::Packed   
)

Definition at line 38 of file matrix-intel.hpp.

References __spv::MatrixA.