|
| 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 () |
|