[PATCH v2 7/7] Support ACEv1 outer product instructions
Haochen Jiang <[email protected]>
| Newsgroups | gmane.comp.gcc.patches |
|---|---|
| Message-ID | <[email protected]> |
Outer product instructions in ACEv1 are quite similar as those VNNI dot product instructions in AVX/AVX512/AVX10 space. Thus, we use the same naming convention as dot product for these insts. gcc/ChangeLog: * config/i386/acev1intrin.h: Add new intrins. * config/i386/i386-builtin-types.def: Add new builtin type. * config/i386/i386-builtin.def (BDESC): Add new builtins. * config/i386/i386-expand.cc (ix86_expand_ace_builtin): Handle new builtin type. * config/i386/sse.md (UNSPECV_TOP2BF16PS): New. (UNSPECV_TOP4BSSD): Ditto. (UNSPECV_TOP4BSSDS): Ditto. (UNSPECV_TOP4BSUD): Ditto. (UNSPECV_TOP4BSUDS): Ditto. (UNSPECV_TOP4BUSD): Ditto. (UNSPECV_TOP4BUSDS): Ditto. (UNSPECV_TOP4BUUD): Ditto. (UNSPECV_TOP4BUUDS): Ditto. (UNSPECV_TOP4MXBF8PS): Ditto. (UNSPECV_TOP4MXBHF8PS): Ditto. (UNSPECV_TOP4MXHBF8PS): Ditto. (UNSPECV_TOP4MXHF8PS): Ditto. (UNSPECV_TOP4MXBSSPS): Ditto. (top2bf16ps): Ditto. (top4<top4bdtype>): Ditto. (top4<top4mxfp8type>ps): Ditto. (top4mxbssps): Ditto. gcc/testsuite/ChangeLog: * gcc.target/i386/acev1-1.c: Add compile test. * gcc.target/i386/ace-check.h: Add helper function. * gcc.target/i386/ace-helper.h: Ditto. * gcc.target/i386/fp8-helper.h: Modify helper function to skip special value check. * gcc.target/i386/avx-1.c: Add acev1 tests. * gcc.target/i386/sse-13.c: Ditto. * gcc.target/i386/sse-14.c: Ditto. * gcc.target/i386/sse-22.c: Ditto. * gcc.target/i386/sse-23.c: Ditto. * gcc.target/i386/acev1-op2bf16ps-2.c: New test. * gcc.target/i386/acev1-op4bssd-2.c: Ditto. * gcc.target/i386/acev1-op4bsud-2.c: Ditto. * gcc.target/i386/acev1-op4busd-2.c: Ditto. * gcc.target/i386/acev1-op4buud-2.c: Ditto. * gcc.target/i386/acev1-op4mxbf8ps-2.c: Ditto. * gcc.target/i386/acev1-op4mxbhf8ps-2.c: Ditto. * gcc.target/i386/acev1-op4mxbssps-2.c: Ditto. * gcc.target/i386/acev1-op4mxhbf8ps-2.c: Ditto. * gcc.target/i386/acev1-op4mxhf8ps-2.c: Ditto. Co-authored-by: Dipesh Sharma <[email protected]> --- gcc/config/i386/acev1intrin.h | 100 ++++++++++++++++++ gcc/config/i386/i386-builtin-types.def | 3 + gcc/config/i386/i386-builtin.def | 10 ++ gcc/config/i386/i386-expand.cc | 18 +++- gcc/config/i386/sse.md | 82 ++++++++++++++ gcc/testsuite/gcc.target/i386/ace-check.h | 21 ++++ gcc/testsuite/gcc.target/i386/ace-helper.h | 11 ++ gcc/testsuite/gcc.target/i386/acev1-1.c | 24 +++++ .../gcc.target/i386/acev1-op2bf16ps-2.c | 47 ++++++++ .../gcc.target/i386/acev1-op4bssd-2.c | 47 ++++++++ .../gcc.target/i386/acev1-op4bsud-2.c | 48 +++++++++ .../gcc.target/i386/acev1-op4busd-2.c | 48 +++++++++ .../gcc.target/i386/acev1-op4buud-2.c | 46 ++++++++ .../gcc.target/i386/acev1-op4mxbf8ps-2.c | 59 +++++++++++ .../gcc.target/i386/acev1-op4mxbhf8ps-2.c | 59 +++++++++++ .../gcc.target/i386/acev1-op4mxbssps-2.c | 54 ++++++++++ .../gcc.target/i386/acev1-op4mxhbf8ps-2.c | 59 +++++++++++ .../gcc.target/i386/acev1-op4mxhf8ps-2.c | 59 +++++++++++ gcc/testsuite/gcc.target/i386/avx-1.c | 10 ++ gcc/testsuite/gcc.target/i386/fp8-helper.h | 21 ++-- gcc/testsuite/gcc.target/i386/sse-13.c | 10 ++ gcc/testsuite/gcc.target/i386/sse-14.c | 14 +++ gcc/testsuite/gcc.target/i386/sse-22.c | 14 +++ gcc/testsuite/gcc.target/i386/sse-23.c | 10 ++ 24 files changed, 862 insertions(+), 12 deletions(-) create mode 100644 gcc/testsuite/gcc.target/i386/acev1-op2bf16ps-2.c create mode 100644 gcc/testsuite/gcc.target/i386/acev1-op4bssd-2.c create mode 100644 gcc/testsuite/gcc.target/i386/acev1-op4bsud-2.c create mode 100644 gcc/testsuite/gcc.target/i386/acev1-op4busd-2.c create mode 100644 gcc/testsuite/gcc.target/i386/acev1-op4buud-2.c create mode 100644 gcc/testsuite/gcc.target/i386/acev1-op4mxbf8ps-2.c create mode 100644 gcc/testsuite/gcc.target/i386/acev1-op4mxbhf8ps-2.c create mode 100644 gcc/testsuite/gcc.target/i386/acev1-op4mxbssps-2.c create mode 100644 gcc/testsuite/gcc.target/i386/acev1-op4mxhbf8ps-2.c create mode 100644 gcc/testsuite/gcc.target/i386/acev1-op4mxhf8ps-2.c diff --git a/gcc/config/i386/acev1intrin.h b/gcc/config/i386/acev1intrin.h index 7b205304302..a449c52af01 100644 --- a/gcc/config/i386/acev1intrin.h +++ b/gcc/config/i386/acev1intrin.h @@ -155,6 +155,76 @@ _tile_insertcol (const int __A, __m512i __B, int __C) __builtin_ia32_tilemovcolinsert (__A, (__v16si) __B, __C); } +extern __inline void +__attribute__((__gnu_inline__, __always_inline__, __artificial__)) +_tile_op2bf16_ps (const int __W, __m512bh __A, __m512bh __B) +{ + __builtin_ia32_top2bf16ps (__W, (__v32bf) __A, (__v32bf) __B); +} + +extern __inline void +__attribute__((__gnu_inline__, __always_inline__, __artificial__)) +_tile_op4bssd_epi32 (const int __W, __m512i __A, __m512i __B) +{ + __builtin_ia32_top4bssd (__W, (__v64qi) __A, (__v64qi) __B); +} + +extern __inline void +__attribute__((__gnu_inline__, __always_inline__, __artificial__)) +_tile_op4bsud_epi32 (const int __W, __m512i __A, __m512i __B) +{ + __builtin_ia32_top4bsud (__W, (__v64qi) __A, (__v64qi) __B); +} + +extern __inline void +__attribute__((__gnu_inline__, __always_inline__, __artificial__)) +_tile_op4busd_epi32 (const int __W, __m512i __A, __m512i __B) +{ + __builtin_ia32_top4busd (__W, (__v64qi) __A, (__v64qi) __B); +} + +extern __inline void +__attribute__((__gnu_inline__, __always_inline__, __artificial__)) +_tile_op4buud_epi32 (const int __W, __m512i __A, __m512i __B) +{ + __builtin_ia32_top4buud (__W, (__v64qi) __A, (__v64qi) __B); +} + +extern __inline void +__attribute__((__gnu_inline__, __always_inline__, __artificial__)) +_tile_op4mxbf8_ps (const int __W, __m512i __A, __m512i __B, const int __C) +{ + __builtin_ia32_top4mxbf8ps (__W, (__v64qi) __A, (__v64qi) __B, __C); +} + +extern __inline void +__attribute__((__gnu_inline__, __always_inline__, __artificial__)) +_tile_op4mxbhf8_ps (const int __W, __m512i __A, __m512i __B, const int __C) +{ + __builtin_ia32_top4mxbhf8ps (__W, (__v64qi) __A, (__v64qi) __B, __C); +} + +extern __inline void +__attribute__((__gnu_inline__, __always_inline__, __artificial__)) +_tile_op4mxhbf8_ps (const int __W, __m512i __A, __m512i __B, const int __C) +{ + __builtin_ia32_top4mxhbf8ps (__W, (__v64qi) __A, (__v64qi) __B, __C); +} + +extern __inline void +__attribute__((__gnu_inline__, __always_inline__, __artificial__)) +_tile_op4mxhf8_ps (const int __W, __m512i __A, __m512i __B, const int __C) +{ + __builtin_ia32_top4mxhf8ps (__W, (__v64qi) __A, (__v64qi) __B, __C); +} + +extern __inline void +__attribute__((__gnu_inline__, __always_inline__, __artificial__)) +_tile_op4mxbss_ps (const int __W, __m512i __A, __m512i __B, const int __C) +{ + __builtin_ia32_top4mxbssps (__W, (__v64qi) __A, (__v64qi) __B, __C); +} + #else #define _tile_ace_zero(A) \ __builtin_ia32_tilezero (A); @@ -183,6 +253,36 @@ _tile_insertcol (const int __A, __m512i __B, int __C) #define _tile_insertcol(A, B, C) \ __builtin_ia32_tilemovcolinsert ((A), (__v16si) (B), (C)) +#define _tile_op2bf16_ps(W, A, B) \ + __builtin_ia32_top2bf16ps (W, (__v32bf) (A), (__v32bf) (B)) + +#define _tile_op4bssd_epi32(W, A, B) \ + __builtin_ia32_top4bssd (W, (__v64qi) (A), (__v64qi) (B)) + +#define _tile_op4bsud_epi32(W, A, B) \ + __builtin_ia32_top4bsud (W, (__v64qi) (A), (__v64qi) (B)) + +#define _tile_op4busd_epi32(W, A, B) \ + __builtin_ia32_top4busd (W, (__v64qi) (A), (__v64qi) (B)) + +#define _tile_op4buud_epi32(W, A, B) \ + __builtin_ia32_top4buud (W, (__v64qi) (A), (__v64qi) (B)) + +#define _tile_op4mxbf8_ps(W, A, B, C) \ + __builtin_ia32_top4mxbf8ps (W, (__v64qi) (A), (__v64qi) (B), C) + +#define _tile_op4mxbhf8_ps(W, A, B, C) \ + __builtin_ia32_top4mxbhf8ps (W, (__v64qi) (A), (__v64qi) (B), C) + +#define _tile_op4mxhbf8_ps(W, A, B, C) \ + __builtin_ia32_top4mxhbf8ps (W, (__v64qi) (A), (__v64qi) (B), C) + +#define _tile_op4mxhf8_ps(W, A, B, C) \ + __builtin_ia32_top4mxhf8ps (W, (__v64qi) (A), (__v64qi) (B), C) + +#define _tile_op4mxbss_ps(W, A, B, C) \ + __builtin_ia32_top4mxbssps (W, (__v64qi) (A), (__v64qi) (B), C) + #endif /* __OPTIMIZE__ */ #endif /* __x86_64__ */ diff --git a/gcc/config/i386/i386-builtin-types.def b/gcc/config/i386/i386-builtin-types.def index 9b969d0d4a9..30a978770eb 100644 --- a/gcc/config/i386/i386-builtin-types.def +++ b/gcc/config/i386/i386-builtin-types.def @@ -1510,3 +1510,6 @@ DEF_FUNCTION_TYPE (V32HF, UQI, SI) DEF_FUNCTION_TYPE (V32BF, UQI, SI) DEF_FUNCTION_TYPE (V16SI, UQI, SI) DEF_FUNCTION_TYPE (VOID, UQI, V16SI, SI) +DEF_FUNCTION_TYPE (VOID, UQI, V32BF, V32BF) +DEF_FUNCTION_TYPE (VOID, UQI, V64QI, V64QI) +DEF_FUNCTION_TYPE (VOID, UQI, V64QI, V64QI, SI) diff --git a/gcc/config/i386/i386-builtin.def b/gcc/config/i386/i386-builtin.def index 50b60e60c9c..e2b7714d4c3 100644 --- a/gcc/config/i386/i386-builtin.def +++ b/gcc/config/i386/i386-builtin.def @@ -3986,5 +3986,15 @@ BDESC (OPTION_MASK_ISA_64BIT, OPTION_MASK_ISA2_ACEV1, CODE_FOR_tcvtrowps2phl, "_ BDESC (OPTION_MASK_ISA_64BIT, OPTION_MASK_ISA2_ACEV1, CODE_FOR_tilemovrow_extract, "__builtin_ia32_tilemovrowextract", IX86_BUILTIN_TILEMOVROWEXTRACT, UNKNOWN, (int) V16SI_FTYPE_UQI_SI) BDESC (OPTION_MASK_ISA_64BIT, OPTION_MASK_ISA2_ACEV1, CODE_FOR_tilemovrow_insert, "__builtin_ia32_tilemovrowinsert", IX86_BUILTIN_TILEMOVROWINSERT, UNKNOWN, (int) VOID_FTYPE_UQI_V16SI_SI) BDESC (OPTION_MASK_ISA_64BIT, OPTION_MASK_ISA2_ACEV1, CODE_FOR_tilemovcol_insert, "__builtin_ia32_tilemovcolinsert", IX86_BUILTIN_TILEMOVCOLINSERT, UNKNOWN, (int) VOID_FTYPE_UQI_V16SI_SI) +BDESC (OPTION_MASK_ISA_64BIT, OPTION_MASK_ISA2_ACEV1, CODE_FOR_top2bf16ps, "__builtin_ia32_top2bf16ps", IX86_BUILTIN_TOP2BF16PS, UNKNOWN, (int) VOID_FTYPE_UQI_V32BF_V32BF) +BDESC (OPTION_MASK_ISA_64BIT, OPTION_MASK_ISA2_ACEV1, CODE_FOR_top4bssd, "__builtin_ia32_top4bssd", IX86_BUILTIN_TOP4BSSD, UNKNOWN, (int) VOID_FTYPE_UQI_V64QI_V64QI) +BDESC (OPTION_MASK_ISA_64BIT, OPTION_MASK_ISA2_ACEV1, CODE_FOR_top4bsud, "__builtin_ia32_top4bsud", IX86_BUILTIN_TOP4BSUD, UNKNOWN, (int) VOID_FTYPE_UQI_V64QI_V64QI) +BDESC (OPTION_MASK_ISA_64BIT, OPTION_MASK_ISA2_ACEV1, CODE_FOR_top4busd, "__builtin_ia32_top4busd", IX86_BUILTIN_TOP4BUSD, UNKNOWN, (int) VOID_FTYPE_UQI_V64QI_V64QI) +BDESC (OPTION_MASK_ISA_64BIT, OPTION_MASK_ISA2_ACEV1, CODE_FOR_top4buud, "__builtin_ia32_top4buud", IX86_BUILTIN_TOP4BUUD, UNKNOWN, (int) VOID_FTYPE_UQI_V64QI_V64QI) +BDESC (OPTION_MASK_ISA_64BIT, OPTION_MASK_ISA2_ACEV1, CODE_FOR_top4mxbf8ps, "__builtin_ia32_top4mxbf8ps", IX86_BUILTIN_TOP4MXBF8PS, UNKNOWN, (int) VOID_FTYPE_UQI_V64QI_V64QI_SI) +BDESC (OPTION_MASK_ISA_64BIT, OPTION_MASK_ISA2_ACEV1, CODE_FOR_top4mxbhf8ps, "__builtin_ia32_top4mxbhf8ps", IX86_BUILTIN_TOP4MXBHF8PS, UNKNOWN, (int) VOID_FTYPE_UQI_V64QI_V64QI_SI) +BDESC (OPTION_MASK_ISA_64BIT, OPTION_MASK_ISA2_ACEV1, CODE_FOR_top4mxhbf8ps, "__builtin_ia32_top4mxhbf8ps", IX86_BUILTIN_TOP4MXHBF8PS, UNKNOWN, (int) VOID_FTYPE_UQI_V64QI_V64QI_SI) +BDESC (OPTION_MASK_ISA_64BIT, OPTION_MASK_ISA2_ACEV1, CODE_FOR_top4mxhf8ps, "__builtin_ia32_top4mxhf8ps", IX86_BUILTIN_TOP4MXHF8PS, UNKNOWN, (int) VOID_FTYPE_UQI_V64QI_V64QI_SI) +BDESC (OPTION_MASK_ISA_64BIT, OPTION_MASK_ISA2_ACEV1, CODE_FOR_top4mxbssps, "__builtin_ia32_top4mxbssps", IX86_BUILTIN_TOP4MXBSSPS, UNKNOWN, (int) VOID_FTYPE_UQI_V64QI_V64QI_SI) BDESC_END (ACE, MAX) diff --git a/gcc/config/i386/i386-expand.cc b/gcc/config/i386/i386-expand.cc index e9bc4a7d8e6..bee8218dae9 100644 --- a/gcc/config/i386/i386-expand.cc +++ b/gcc/config/i386/i386-expand.cc @@ -14716,7 +14716,7 @@ ix86_expand_ace_builtin (const struct builtin_description *d, tree exp, { tree arg; rtx pat, op; - unsigned int i, nargs, arg_adjust = 0; + unsigned int i, nargs, arg_adjust = 0, constant = 100; bool tmm_src = false; rtx xops[4]; enum insn_code icode = d->icode; @@ -14735,8 +14735,14 @@ ix86_expand_ace_builtin (const struct builtin_description *d, tree exp, tmm_src = true; break; case VOID_FTYPE_UQI_V16SI_SI: + case VOID_FTYPE_UQI_V32BF_V32BF: + case VOID_FTYPE_UQI_V64QI_V64QI: nargs = 3; break; + case VOID_FTYPE_UQI_V64QI_V64QI_SI: + nargs = 4; + constant = 3; + break; default: gcc_unreachable (); @@ -14762,16 +14768,17 @@ ix86_expand_ace_builtin (const struct builtin_description *d, tree exp, arg = CALL_EXPR_ARG (exp, i); op = ix86_expand_unsigned_small_int_cst_argument (arg); - if (i == 0) + if (i == 0 || i == constant) { - /* This must be the tmm reg number constant. */ + /* This must be the constant. */ if (!insn_p->operand[i + arg_adjust].predicate(op, SImode)) { error ("the argument must be constant"); return const0_rtx; } - if (!IN_RANGE (INTVAL (op), 0, 7)) + /* This must be the tmm reg number constant. */ + if (i == 0 && !IN_RANGE (INTVAL (op), 0, 7)) { error ("the tmm register number argument must be between 0 to 7"); return const0_rtx; @@ -14819,6 +14826,9 @@ ix86_expand_ace_builtin (const struct builtin_description *d, tree exp, case 3: pat = GEN_FCN (icode) (xops[0], xops[1], xops[2]); break; + case 4: + pat = GEN_FCN (icode) (xops[0], xops[1], xops[2], xops[3]); + break; default: gcc_unreachable (); } diff --git a/gcc/config/i386/sse.md b/gcc/config/i386/sse.md index 6f3bf44b557..0a3ab0d29cc 100644 --- a/gcc/config/i386/sse.md +++ b/gcc/config/i386/sse.md @@ -320,6 +320,20 @@ UNSPECV_BSRMOVL_LOAD UNSPECV_TILEMOVROWINSERT UNSPECV_TILEMOVCOLINSERT + UNSPECV_TOP2BF16PS + UNSPECV_TOP4BSSD + UNSPECV_TOP4BSSDS + UNSPECV_TOP4BSUD + UNSPECV_TOP4BSUDS + UNSPECV_TOP4BUSD + UNSPECV_TOP4BUSDS + UNSPECV_TOP4BUUD + UNSPECV_TOP4BUUDS + UNSPECV_TOP4MXBF8PS + UNSPECV_TOP4MXBHF8PS + UNSPECV_TOP4MXHBF8PS + UNSPECV_TOP4MXHF8PS + UNSPECV_TOP4MXBSSPS ]) ;; All vector modes including V?TImode, used in move patterns. @@ -35058,3 +35072,71 @@ "TARGET_ACEV1" "tilemov<rowcol>\t{%2, %1, %%tmm%c0|tmm%c0, %1, %2}" [(set_attr "prefix" "evex")]) + +(define_insn "top2bf16ps" + [(set (reg:V32SF TMM_REGNUM) + (unspec_volatile:V32SF + [(match_operand:QI 0 "const_0_to_7_operand") + (match_operand:V32BF 1 "register_operand" "v") + (match_operand:V32BF 2 "register_operand" "v")] + UNSPECV_TOP2BF16PS))] + "TARGET_ACEV1" + "top2bf16ps\t{%2, %1, %%tmm%c0|tmm%c0, %1, %2}" + [(set_attr "prefix" "evex")]) + +(define_int_iterator TOP4BDTYPE + [UNSPECV_TOP4BSSD + UNSPECV_TOP4BSUD + UNSPECV_TOP4BUSD + UNSPECV_TOP4BUUD]) + +(define_int_attr top4bdtype + [(UNSPECV_TOP4BSSD "bssd") (UNSPECV_TOP4BSUD "bsud") + (UNSPECV_TOP4BUSD "busd") (UNSPECV_TOP4BUUD "buud")]) + +(define_insn "top4<top4bdtype>" + [(set (reg:V32SI TMM_REGNUM) + (unspec_volatile:V32SI + [(match_operand:QI 0 "const_0_to_7_operand") + (match_operand:V64QI 1 "register_operand" "v") + (match_operand:V64QI 2 "register_operand" "v")] + TOP4BDTYPE))] + "TARGET_ACEV1" + "top4<top4bdtype>\t{%2, %1, %%tmm%c0|tmm%c0, %1, %2}" + [(set_attr "prefix" "evex")]) + +(define_int_iterator TOP4MXFP8TYPE + [UNSPECV_TOP4MXBF8PS UNSPECV_TOP4MXBHF8PS + UNSPECV_TOP4MXHBF8PS UNSPECV_TOP4MXHF8PS]) + +(define_int_attr top4mxfp8type + [(UNSPECV_TOP4MXBF8PS "mxbf8") + (UNSPECV_TOP4MXBHF8PS "mxbhf8") + (UNSPECV_TOP4MXHBF8PS "mxhbf8") + (UNSPECV_TOP4MXHF8PS "mxhf8")]) + +(define_insn "top4<top4mxfp8type>ps" + [(set (reg:V32SF TMM_REGNUM) + (unspec_volatile:V32SF + [(match_operand:QI 0 "const_0_to_7_operand") + (match_operand:V64QI 1 "register_operand" "v") + (match_operand:V64QI 2 "register_operand" "v") + (match_operand:SI 3 "const_0_to_255_operand") + (reg:V32SI BSR0_REG)] + TOP4MXFP8TYPE))] + "TARGET_ACEV1" + "top4<top4mxfp8type>ps\t{%3, %2, %1, %%tmm%c0|tmm%c0, %1, %2, %3}" + [(set_attr "prefix" "evex")]) + +(define_insn "top4mxbssps" + [(set (reg:V32SF TMM_REGNUM) + (unspec_volatile:V32SF + [(match_operand:QI 0 "const_0_to_7_operand") + (match_operand:V64QI 1 "register_operand" "v") + (match_operand:V64QI 2 "register_operand" "v") + (match_operand:SI 3 "const_0_to_255_operand") + (reg:V32SI BSR0_REG)] + UNSPECV_TOP4MXBSSPS))] + "TARGET_ACEV1" + "top4mxbssps\t{%3, %2, %1, %%tmm%c0|tmm%c0, %1, %2, %3}" + [(set_attr "prefix" "evex")]) diff --git a/gcc/testsuite/gcc.target/i386/ace-check.h b/gcc/testsuite/gcc.target/i386/ace-check.h index 210f20e7a47..e153c5ff84b 100644 --- a/gcc/testsuite/gcc.target/i386/ace-check.h +++ b/gcc/testsuite/gcc.target/i386/ace-check.h @@ -63,6 +63,27 @@ void init_tile_config (__tilecfg *dst, __bsr* bsr) _bsr0_init (); } +#define CHECK_TILE_REGISTER(regnum, dst_ref) \ +{ \ + int miss = 0; \ + for (int j = 0; j < 16; j++) \ + { \ + __m512i tmp; \ + union512i_ub a; \ + tmp = _tile_extractrow (regnum, j); \ + a.x = tmp; \ + for (int k = 0; k < 64; k++) \ + if (a.a[k] != dst_ref.buf[j * 64 + k]) \ + { \ + printf ("Row %d Column %d: 0x%x != 0x%x\n", j, k, a.a[k], \ + dst_ref.buf[j * 64 + k]); \ + miss += 1; \ + } \ + } \ + if (miss) \ + abort (); \ +} + #ifndef DO_TEST #define DO_TEST do_test static void test_ace (void); diff --git a/gcc/testsuite/gcc.target/i386/ace-helper.h b/gcc/testsuite/gcc.target/i386/ace-helper.h index 8b4611e6e3f..9f9401aa2b2 100644 --- a/gcc/testsuite/gcc.target/i386/ace-helper.h +++ b/gcc/testsuite/gcc.target/i386/ace-helper.h @@ -4,4 +4,15 @@ #define AVX512FP16 #define AVX512BF16 #include "avx512f-helper.h" +#include "fp8-helper.h" + +float convert_e8m0_to_fp32 (unsigned char x) +{ + unsigned int tmp = ((unsigned int) x) << 23; + Floatuint32Union ux = { .u = tmp }; + if (x == 0xff) + ux.u |= 0x1; + return ux.f; +} + #endif diff --git a/gcc/testsuite/gcc.target/i386/acev1-1.c b/gcc/testsuite/gcc.target/i386/acev1-1.c index 6d01745c592..8976b555a4a 100644 --- a/gcc/testsuite/gcc.target/i386/acev1-1.c +++ b/gcc/testsuite/gcc.target/i386/acev1-1.c @@ -15,6 +15,16 @@ /* { dg-final { scan-assembler-times "tcvtrowps2phl\[ \t]" 1 } } */ /* { dg-final { scan-assembler-times "tilemovrow\[ \t]" 2 } } */ /* { dg-final { scan-assembler-times "tilemovcol\[ \t]" 1 } } */ +/* { dg-final { scan-assembler-times "top2bf16ps\[ \t]" 1 } } */ +/* { dg-final { scan-assembler-times "top4bssd\[ \t]" 1 } } */ +/* { dg-final { scan-assembler-times "top4bsud\[ \t]" 1 } } */ +/* { dg-final { scan-assembler-times "top4busd\[ \t]" 1 } } */ +/* { dg-final { scan-assembler-times "top4buud\[ \t]" 1 } } */ +/* { dg-final { scan-assembler-times "top4mxbf8ps\[ \t]" 1 } } */ +/* { dg-final { scan-assembler-times "top4mxbhf8ps\[ \t]" 1 } } */ +/* { dg-final { scan-assembler-times "top4mxhbf8ps\[ \t]" 1 } } */ +/* { dg-final { scan-assembler-times "top4mxhf8ps\[ \t]" 1 } } */ +/* { dg-final { scan-assembler-times "top4mxbssps\[ \t]" 1 } } */ #include <immintrin.h> extern int t[]; @@ -52,3 +62,17 @@ void cvtrow () _tile_insertrow (7, a1, 10); _tile_insertcol (2, a2, 11); } + +void op () +{ + _tile_op2bf16_ps (1, b1, b2); + _tile_op4bssd_epi32 (2, a1, a2); + _tile_op4bsud_epi32 (3, a1, a2); + _tile_op4busd_epi32 (4, a1, a2); + _tile_op4buud_epi32 (5, a1, a2); + _tile_op4mxbf8_ps (6, a1, a2, 1); + _tile_op4mxbhf8_ps (7, a1, a2, 6); + _tile_op4mxhbf8_ps (0, a1, a2, 9); + _tile_op4mxhf8_ps (1, a1, a2, 3); + _tile_op4mxbss_ps (2, a1, a2, 4); +} diff --git a/gcc/testsuite/gcc.target/i386/acev1-op2bf16ps-2.c b/gcc/testsuite/gcc.target/i386/acev1-op2bf16ps-2.c new file mode 100644 index 00000000000..c644eea4b77 --- /dev/null +++ b/gcc/testsuite/gcc.target/i386/acev1-op2bf16ps-2.c @@ -0,0 +1,47 @@ +/* { dg-do run { target { ! ia32 } } } */ +/* { dg-require-effective-target acev1 } */ +/* { dg-options "-O2 -macev1" } */ +#define DO_TEST test_acev1_op2bf16ps +void test_acev1_op2bf16ps (); +#include "ace-helper.h" + +void calc_op2bf16ps (__tile *dst, __bf16 *src1, __bf16 *src2) +{ + int i, j; + + for (i = 0; i < 16; i++) + for (j = 0; j < 16; j++) + { + dst->a[i * 16 + j] += (float) src1[2 * i] * (float) src2[2 * j]; + dst->a[i * 16 + j] += (float) src1[2 * i + 1] * (float) src2[2 * j + 1]; + } +} + +void test_acev1_op2bf16ps () +{ + __tilecfg cfg; + __tile dst_ref; + __bsr bsr0; + union512bf16_bf src1, src2; + int m; + + for (m = 0; m < 32; m++) + { + int sign; + sign = m % 2 ? 1 : -1; + src1.a[m] = (__bf16) (sign * (1.5 * (1 << (m % 3)))); + sign = m % 5 ? 1 : -1; + src2.a[m] = (__bf16) (sign * (2.5 * (1 << (m % 3)))); + } + + for (m = 0; m < 1024; m++) + dst_ref.buf[m] = 0; + + calc_op2bf16ps (&dst_ref, src1.a, src2.a); + + init_tile_config (&cfg, &bsr0); + _tile_ace_zero (1); + _tile_op2bf16_ps (1, src1.x, src2.x); + + CHECK_TILE_REGISTER (1, dst_ref); +} diff --git a/gcc/testsuite/gcc.target/i386/acev1-op4bssd-2.c b/gcc/testsuite/gcc.target/i386/acev1-op4bssd-2.c new file mode 100644 index 00000000000..5243ff1c931 --- /dev/null +++ b/gcc/testsuite/gcc.target/i386/acev1-op4bssd-2.c @@ -0,0 +1,47 @@ +/* { dg-do run { target { ! ia32 } } } */ +/* { dg-require-effective-target acev1 } */ +/* { dg-options "-O2 -macev1" } */ +#define DO_TEST test_acev1_op4bssd +void test_acev1_op4bssd (); +#include "ace-helper.h" + +void calc_op4bssd (__tile *dst, char *src1, char *src2) +{ + int i, j; + + for (i = 0; i < 16; i++) + for (j = 0; j < 16; j++) + { + dst->b[i * 16 + j] += (int) src1[4 * i] * (int) src2[4 * j]; + dst->b[i * 16 + j] += (int) src1[4 * i + 1] * (int) src2[4 * j + 1]; + dst->b[i * 16 + j] += (int) src1[4 * i + 2] * (int) src2[4 * j + 2]; + dst->b[i * 16 + j] += (int) src1[4 * i + 3] * (int) src2[4 * j + 3]; + } +} + +void test_acev1_op4bssd () +{ + __tilecfg cfg; + __tile dst_ref; + __bsr bsr0; + union512i_b src1, src2; + int m; + + for (m = 0; m < 64; m++) + { + int sign = m % 2 ? 1 : -1; + src1.a[m] = 10 + 3 * m + sign; + src2.a[m] = sign * 10 * m * m; + } + + for (m = 0; m < 1024; m++) + dst_ref.buf[m] = 0; + + calc_op4bssd (&dst_ref, src1.a, src2.a); + + init_tile_config (&cfg, &bsr0); + _tile_ace_zero (1); + _tile_op4bssd_epi32 (1, src1.x, src2.x); + + CHECK_TILE_REGISTER (1, dst_ref); +} diff --git a/gcc/testsuite/gcc.target/i386/acev1-op4bsud-2.c b/gcc/testsuite/gcc.target/i386/acev1-op4bsud-2.c new file mode 100644 index 00000000000..1fed75a61b2 --- /dev/null +++ b/gcc/testsuite/gcc.target/i386/acev1-op4bsud-2.c @@ -0,0 +1,48 @@ +/* { dg-do run { target { ! ia32 } } } */ +/* { dg-require-effective-target acev1 } */ +/* { dg-options "-O2 -macev1" } */ +#define DO_TEST test_acev1_op4bsud +void test_acev1_op4bsud (); +#include "ace-helper.h" + +void calc_op4bsud (__tile *dst, char *src1, unsigned char *src2) +{ + int i, j; + + for (i = 0; i < 16; i++) + for (j = 0; j < 16; j++) + { + dst->b[i * 16 + j] += (int) src1[4 * i] * (int) src2[4 * j]; + dst->b[i * 16 + j] += (int) src1[4 * i + 1] * (int) src2[4 * j + 1]; + dst->b[i * 16 + j] += (int) src1[4 * i + 2] * (int) src2[4 * j + 2]; + dst->b[i * 16 + j] += (int) src1[4 * i + 3] * (int) src2[4 * j + 3]; + } +} + +void test_acev1_op4bsud () +{ + __tilecfg cfg; + __tile dst_ref; + __bsr bsr0; + union512i_b src1; + union512i_ub src2; + int m; + + for (m = 0; m < 64; m++) + { + int sign = m % 2 ? 1 : -1; + src1.a[m] = sign * (10 + 3 * m + sign); + src2.a[m] = 10 * m * m; + } + + for (m = 0; m < 1024; m++) + dst_ref.buf[m] = 0; + + calc_op4bsud (&dst_ref, src1.a, src2.a); + + init_tile_config (&cfg, &bsr0); + _tile_ace_zero (1); + _tile_op4bsud_epi32 (1, src1.x, src2.x); + + CHECK_TILE_REGISTER (1, dst_ref); +} diff --git a/gcc/testsuite/gcc.target/i386/acev1-op4busd-2.c b/gcc/testsuite/gcc.target/i386/acev1-op4busd-2.c new file mode 100644 index 00000000000..236e073f68f --- /dev/null +++ b/gcc/testsuite/gcc.target/i386/acev1-op4busd-2.c @@ -0,0 +1,48 @@ +/* { dg-do run { target { ! ia32 } } } */ +/* { dg-require-effective-target acev1 } */ +/* { dg-options "-O2 -macev1" } */ +#define DO_TEST test_acev1_op4busd +void test_acev1_op4busd (); +#include "ace-helper.h" + +void calc_op4busd (__tile *dst, unsigned char *src1, char *src2) +{ + int i, j; + + for (i = 0; i < 16; i++) + for (j = 0; j < 16; j++) + { + dst->b[i * 16 + j] += (int) src1[4 * i] * (int) src2[4 * j]; + dst->b[i * 16 + j] += (int) src1[4 * i + 1] * (int) src2[4 * j + 1]; + dst->b[i * 16 + j] += (int) src1[4 * i + 2] * (int) src2[4 * j + 2]; + dst->b[i * 16 + j] += (int) src1[4 * i + 3] * (int) src2[4 * j + 3]; + } +} + +void test_acev1_op4busd () +{ + __tilecfg cfg; + __tile dst_ref; + __bsr bsr0; + union512i_ub src1; + union512i_b src2; + int m; + + for (m = 0; m < 64; m++) + { + int sign = m % 2 ? 1 : -1; + src1.a[m] = 10 + 3 * m; + src2.a[m] = sign * 10 * m * m; + } + + for (m = 0; m < 1024; m++) + dst_ref.buf[m] = 0; + + calc_op4busd (&dst_ref, src1.a, src2.a); + + init_tile_config (&cfg, &bsr0); + _tile_ace_zero (1); + _tile_op4busd_epi32 (1, src1.x, src2.x); + + CHECK_TILE_REGISTER (1, dst_ref); +} diff --git a/gcc/testsuite/gcc.target/i386/acev1-op4buud-2.c b/gcc/testsuite/gcc.target/i386/acev1-op4buud-2.c new file mode 100644 index 00000000000..debf48d7958 --- /dev/null +++ b/gcc/testsuite/gcc.target/i386/acev1-op4buud-2.c @@ -0,0 +1,46 @@ +/* { dg-do run { target { ! ia32 } } } */ +/* { dg-require-effective-target acev1 } */ +/* { dg-options "-O2 -macev1" } */ +#define DO_TEST test_acev1_op4buud +void test_acev1_op4buud (); +#include "ace-helper.h" + +void calc_op4buud (__tile *dst, unsigned char *src1, unsigned char *src2) +{ + int i, j; + + for (i = 0; i < 16; i++) + for (j = 0; j < 16; j++) + { + dst->b[i * 16 + j] += (int) src1[4 * i] * (int) src2[4 * j]; + dst->b[i * 16 + j] += (int) src1[4 * i + 1] * (int) src2[4 * j + 1]; + dst->b[i * 16 + j] += (int) src1[4 * i + 2] * (int) src2[4 * j + 2]; + dst->b[i * 16 + j] += (int) src1[4 * i + 3] * (int) src2[4 * j + 3]; + } +} + +void test_acev1_op4buud () +{ + __tilecfg cfg; + __tile dst_ref; + __bsr bsr0; + union512i_ub src1, src2; + int m, n; + + for (m = 0; m < 64; m++) + { + src1.a[m] = 10 + 3 * m; + src2.a[m] = 10 * m * m; + } + + for (m = 0; m < 1024; m++) + dst_ref.buf[m] = 0; + + calc_op4buud (&dst_ref, src1.a, src2.a); + + init_tile_config (&cfg, &bsr0); + _tile_ace_zero (1); + _tile_op4buud_epi32 (1, src1.x, src2.x); + + CHECK_TILE_REGISTER (1, dst_ref); +} diff --git a/gcc/testsuite/gcc.target/i386/acev1-op4mxbf8ps-2.c b/gcc/testsuite/gcc.target/i386/acev1-op4mxbf8ps-2.c new file mode 100644 index 00000000000..e4cfa57a272 --- /dev/null +++ b/gcc/testsuite/gcc.target/i386/acev1-op4mxbf8ps-2.c @@ -0,0 +1,59 @@ +/* { dg-do run { target { ! ia32 } } } */ +/* { dg-require-effective-target acev1 } */ +/* { dg-options "-O2 -macev1" } */ +#define DO_TEST test_acev1_op4mxbf8ps +void test_acev1_op4mxbf8ps (); +#include "ace-helper.h" + +void calc_op4mxbf8ps (__tile *dst, char *src1, char *src2, __bsr* bsr, const int imm) +{ + int i, j; + + for (i = 0; i < 16; i++) + { + for (j = 0; j < 16; j++) + { + __int128_t tmp = 0; + float scale = convert_e8m0_to_fp32 (bsr->buf[4 * i + (imm & 0x3)]) + * convert_e8m0_to_fp32 (bsr->buf[64 + 4 * j + ((imm & 0x18) >> 3)]); + tmp += (__int128_t) shift_fp8_to_int64 (src1[4 * i], 1, NULL) + * (__int128_t) shift_fp8_to_int64 (src2[4 * j], 1, NULL); + tmp += (__int128_t) shift_fp8_to_int64 (src1[4 * i + 1], 1, NULL) + * (__int128_t) shift_fp8_to_int64 (src2[4 * j + 1], 1, NULL); + tmp += (__int128_t) shift_fp8_to_int64 (src1[4 * i + 2], 1, NULL) + * (__int128_t) shift_fp8_to_int64 (src2[4 * j + 2], 1, NULL); + tmp += (__int128_t) shift_fp8_to_int64 (src1[4 * i + 3], 1, NULL) + * (__int128_t) shift_fp8_to_int64 (src2[4 * j + 3], 1, NULL); + dst->a[16 * i + j] += shift_int128_to_fp32 (tmp, 1, 1) * scale; + } + } +} + +void test_acev1_op4mxbf8ps () +{ + __tilecfg cfg; + __tile dst_ref; + __bsr bsr0; + union512i_b src1, src2; + union512i_ub bsrl, bsrh; + int m; + + for (m = 0; m < 64; m++) + { + src1.a[m] = (m % 2) << 7 | (m % 30 + 1) << 2 | ((m + 1) % 4); + src2.a[m] = ((m % 3) % 2) << 7 | ((m + 10) % 30 + 1) << 2 | ((m + 2) % 4); + } + + for (m = 0; m < 1024; m++) + dst_ref.buf[m] = 0; + + init_tile_config (&cfg, &bsr0); + fill_bsr (&bsr0, &bsrl, &bsrh); + calc_op4mxbf8ps (&dst_ref, src1.a, src2.a, &bsr0, 10); + + _tile_ace_zero (1); + _bsr0_insertfull (bsrh.x, bsrl.x); + _tile_op4mxbf8_ps (1, src1.x, src2.x, 10); + + CHECK_TILE_REGISTER (1, dst_ref); +} diff --git a/gcc/testsuite/gcc.target/i386/acev1-op4mxbhf8ps-2.c b/gcc/testsuite/gcc.target/i386/acev1-op4mxbhf8ps-2.c new file mode 100644 index 00000000000..6ef4f7113e4 --- /dev/null +++ b/gcc/testsuite/gcc.target/i386/acev1-op4mxbhf8ps-2.c @@ -0,0 +1,59 @@ +/* { dg-do run { target { ! ia32 } } } */ +/* { dg-require-effective-target acev1 } */ +/* { dg-options "-O2 -macev1" } */ +#define DO_TEST test_acev1_op4mxbhf8ps +void test_acev1_op4mxbhf8ps (); +#include "ace-helper.h" + +void calc_op4mxbhf8ps (__tile *dst, char *src1, char *src2, __bsr* bsr, const int imm) +{ + int i, j; + + for (i = 0; i < 16; i++) + { + for (j = 0; j < 16; j++) + { + __int128_t tmp = 0; + float scale = convert_e8m0_to_fp32 (bsr->buf[4 * i + (imm & 0x3)]) + * convert_e8m0_to_fp32 (bsr->buf[64 + 4 * j + ((imm & 0x18) >> 3)]); + tmp += (__int128_t) shift_fp8_to_int64 (src1[4 * i], 1, NULL) + * (__int128_t) shift_fp8_to_int64 (src2[4 * j], 0, NULL); + tmp += (__int128_t) shift_fp8_to_int64 (src1[4 * i + 1], 1, NULL) + * (__int128_t) shift_fp8_to_int64 (src2[4 * j + 1], 0, NULL); + tmp += (__int128_t) shift_fp8_to_int64 (src1[4 * i + 2], 1, NULL) + * (__int128_t) shift_fp8_to_int64 (src2[4 * j + 2], 0, NULL); + tmp += (__int128_t) shift_fp8_to_int64 (src1[4 * i + 3], 1, NULL) + * (__int128_t) shift_fp8_to_int64 (src2[4 * j + 3], 0, NULL); + dst->a[16 * i + j] += shift_int128_to_fp32 (tmp, 1, 0) * scale; + } + } +} + +void test_acev1_op4mxbhf8ps () +{ + __tilecfg cfg; + __tile dst_ref; + __bsr bsr0; + union512i_b src1, src2; + union512i_ub bsrl, bsrh; + int m; + + for (m = 0; m < 64; m++) + { + src1.a[m] = (m % 2) << 7 | (m % 31) << 2 | ((m + 1) % 4); + src2.a[m] = ((m % 3) % 2) << 7 | ((m + 7) % 16) << 3 | ((m + 2) % 8); + } + + for (m = 0; m < 1024; m++) + dst_ref.buf[m] = 0; + + init_tile_config (&cfg, &bsr0); + fill_bsr (&bsr0, &bsrl, &bsrh); + calc_op4mxbhf8ps (&dst_ref, src1.a, src2.a, &bsr0, 1); + + _tile_ace_zero (1); + _bsr0_insertfull (bsrh.x, bsrl.x); + _tile_op4mxbhf8_ps (1, src1.x, src2.x, 1); + + CHECK_TILE_REGISTER (1, dst_ref); +} diff --git a/gcc/testsuite/gcc.target/i386/acev1-op4mxbssps-2.c b/gcc/testsuite/gcc.target/i386/acev1-op4mxbssps-2.c new file mode 100644 index 00000000000..0fe5a3a7e8a --- /dev/null +++ b/gcc/testsuite/gcc.target/i386/acev1-op4mxbssps-2.c @@ -0,0 +1,54 @@ +/* { dg-do run { target { ! ia32 } } } */ +/* { dg-require-effective-target acev1 } */ +/* { dg-options "-O2 -macev1" } */ +#define DO_TEST test_acev1_op4mxbssps +void test_acev1_op4mxbssps (); +#include "ace-helper.h" + +void calc_op4mxbssps (__tile *dst, char *src1, char *src2, __bsr* bsr, const int imm) +{ + int i, j; + + for (i = 0; i < 16; i++) + for (j = 0; j < 16; j++) + { + int tmp = 0; + float scale = convert_e8m0_to_fp32 (bsr->buf[4 * i + (imm & 0x3)]) + * convert_e8m0_to_fp32 (bsr->buf[64 + 4 * j + ((imm & 0x18) >> 3)]); + tmp += (int) src1[4 * i] * (int) src2[4 * j]; + tmp += (int) src1[4 * i + 1] * (int) src2[4 * j + 1]; + tmp += (int) src1[4 * i + 2] * (int) src2[4 * j + 2]; + tmp += (int) src1[4 * i + 3] * (int) src2[4 * j + 3]; + dst->a[16 * i + j] += (float) tmp * scale / 4096.0; + } +} + +void test_acev1_op4mxbssps () +{ + __tilecfg cfg; + __tile dst_ref; + __bsr bsr0; + union512i_b src1, src2; + union512i_ub bsrl, bsrh; + int m; + + for (m = 0; m < 64; m++) + { + int sign = m % 2 ? 1 : -1; + src1.a[m] = 10 + 3 * m + sign; + src2.a[m] = sign * 10 * m * m; + } + + for (m = 0; m < 1024; m++) + dst_ref.buf[m] = 0; + + init_tile_config (&cfg, &bsr0); + fill_bsr (&bsr0, &bsrl, &bsrh); + calc_op4mxbssps (&dst_ref, src1.a, src2.a, &bsr0, 10); + + _tile_ace_zero (1); + _bsr0_insertfull (bsrh.x, bsrl.x); + _tile_op4mxbss_ps (1, src1.x, src2.x, 10); + + CHECK_TILE_REGISTER (1, dst_ref); +} diff --git a/gcc/testsuite/gcc.target/i386/acev1-op4mxhbf8ps-2.c b/gcc/testsuite/gcc.target/i386/acev1-op4mxhbf8ps-2.c new file mode 100644 index 00000000000..350f108b71c --- /dev/null +++ b/gcc/testsuite/gcc.target/i386/acev1-op4mxhbf8ps-2.c @@ -0,0 +1,59 @@ +/* { dg-do run { target { ! ia32 } } } */ +/* { dg-require-effective-target acev1 } */ +/* { dg-options "-O2 -macev1" } */ +#define DO_TEST test_acev1_op4mxhbf8ps +void test_acev1_op4mxhbf8ps (); +#include "ace-helper.h" + +void calc_op4mxhbf8ps (__tile *dst, char *src1, char *src2, __bsr* bsr, const int imm) +{ + int i, j; + + for (i = 0; i < 16; i++) + { + for (j = 0; j < 16; j++) + { + __int128_t tmp = 0; + float scale = convert_e8m0_to_fp32 (bsr->buf[4 * i + (imm & 0x3)]) + * convert_e8m0_to_fp32 (bsr->buf[64 + 4 * j + ((imm & 0x18) >> 3)]); + tmp += (__int128_t) shift_fp8_to_int64 (src1[4 * i], 0, NULL) + * (__int128_t) shift_fp8_to_int64 (src2[4 * j], 1, NULL); + tmp += (__int128_t) shift_fp8_to_int64 (src1[4 * i + 1], 0, NULL) + * (__int128_t) shift_fp8_to_int64 (src2[4 * j + 1], 1, NULL); + tmp += (__int128_t) shift_fp8_to_int64 (src1[4 * i + 2], 0, NULL) + * (__int128_t) shift_fp8_to_int64 (src2[4 * j + 2], 1, NULL); + tmp += (__int128_t) shift_fp8_to_int64 (src1[4 * i + 3], 0, NULL) + * (__int128_t) shift_fp8_to_int64 (src2[4 * j + 3], 1, NULL); + dst->a[16 * i + j] += shift_int128_to_fp32 (tmp, 0, 1) * scale; + } + } +} + +void test_acev1_op4mxhbf8ps () +{ + __tilecfg cfg; + __tile dst_ref; + __bsr bsr0; + union512i_b src1, src2; + union512i_ub bsrl, bsrh; + int m; + + for (m = 0; m < 64; m++) + { + src1.a[m] = (m % 2) << 7 | (m % 16) << 3 | ((m + 1) % 8); + src2.a[m] = ((m % 3) % 2) << 7 | ((m + 10) % 31) << 2| ((m + 2) % 4); + } + + for (m = 0; m < 1024; m++) + dst_ref.buf[m] = 0; + + init_tile_config (&cfg, &bsr0); + fill_bsr (&bsr0, &bsrl, &bsrh); + calc_op4mxhbf8ps (&dst_ref, src1.a, src2.a, &bsr0, 17); + + _tile_ace_zero (1); + _bsr0_insertfull (bsrh.x, bsrl.x); + _tile_op4mxhbf8_ps (1, src1.x, src2.x, 17); + + CHECK_TILE_REGISTER (1, dst_ref); +} diff --git a/gcc/testsuite/gcc.target/i386/acev1-op4mxhf8ps-2.c b/gcc/testsuite/gcc.target/i386/acev1-op4mxhf8ps-2.c new file mode 100644 index 00000000000..c3329026b11 --- /dev/null +++ b/gcc/testsuite/gcc.target/i386/acev1-op4mxhf8ps-2.c @@ -0,0 +1,59 @@ +/* { dg-do run { target { ! ia32 } } } */ +/* { dg-require-effective-target acev1 } */ +/* { dg-options "-O2 -macev1" } */ +#define DO_TEST test_acev1_op4mxhf8ps +void test_acev1_op4mxhf8ps (); +#include "ace-helper.h" + +void calc_op4mxhf8ps (__tile *dst, char *src1, char *src2, __bsr* bsr, const int imm) +{ + int i, j; + + for (i = 0; i < 16; i++) + { + for (j = 0; j < 16; j++) + { + __int128_t tmp = 0; + float scale = convert_e8m0_to_fp32 (bsr->buf[4 * i + (imm & 0x3)]) + * convert_e8m0_to_fp32 (bsr->buf[64 + 4 * j + ((imm & 0x18) >> 3)]); + tmp += (__int128_t) shift_fp8_to_int64 (src1[4 * i], 0, NULL) + * (__int128_t) shift_fp8_to_int64 (src2[4 * j], 0, NULL); + tmp += (__int128_t) shift_fp8_to_int64 (src1[4 * i + 1], 0, NULL) + * (__int128_t) shift_fp8_to_int64 (src2[4 * j + 1], 0, NULL); + tmp += (__int128_t) shift_fp8_to_int64 (src1[4 * i + 2], 0, NULL) + * (__int128_t) shift_fp8_to_int64 (src2[4 * j + 2], 0, NULL); + tmp += (__int128_t) shift_fp8_to_int64 (src1[4 * i + 3], 0, NULL) + * (__int128_t) shift_fp8_to_int64 (src2[4 * j + 3], 0, NULL); + dst->a[16 * i + j] += shift_int128_to_fp32 (tmp, 0, 0) * scale; + } + } +} + +void test_acev1_op4mxhf8ps () +{ + __tilecfg cfg; + __tile dst_ref; + __bsr bsr0; + union512i_b src1, src2; + union512i_ub bsrl, bsrh; + int m; + + for (m = 0; m < 64; m++) + { + src1.a[m] = (m % 2) << 7 | (m % 16) << 3 | ((m + 1) % 8); + src2.a[m] = ((m % 3) % 2) << 7 | ((m + 7) % 16) << 3 | ((m + 2) % 8); + } + + for (m = 0; m < 1024; m++) + dst_ref.buf[m] = 0; + + init_tile_config (&cfg, &bsr0); + fill_bsr (&bsr0, &bsrl, &bsrh); + calc_op4mxhf8ps (&dst_ref, src1.a, src2.a, &bsr0, 16); + + _tile_ace_zero (1); + _bsr0_insertfull (bsrh.x, bsrl.x); + _tile_op4mxhf8_ps (1, src1.x, src2.x, 16); + + CHECK_TILE_REGISTER (1, dst_ref); +} diff --git a/gcc/testsuite/gcc.target/i386/avx-1.c b/gcc/testsuite/gcc.target/i386/avx-1.c index 623446deec8..7a8e5cc2301 100644 --- a/gcc/testsuite/gcc.target/i386/avx-1.c +++ b/gcc/testsuite/gcc.target/i386/avx-1.c @@ -929,6 +929,16 @@ #define __builtin_ia32_tilemovrowextract(A, B) __builtin_ia32_tilemovrowextract (1, B) #define __builtin_ia32_tilemovrowinsert(A, B, C) __builtin_ia32_tilemovrowinsert (1, B, C) #define __builtin_ia32_tilemovcolinsert(A, B, C) __builtin_ia32_tilemovcolinsert (1, B, C) +#define __builtin_ia32_top2bf16ps(W, A, B) __builtin_ia32_top2bf16ps (1, A, B) +#define __builtin_ia32_top4bssd(W, A, B) __builtin_ia32_top4bssd (1, A, B) +#define __builtin_ia32_top4bsud(W, A, B) __builtin_ia32_top4bsud (1, A, B) +#define __builtin_ia32_top4busd(W, A, B) __builtin_ia32_top4busd (1, A, B) +#define __builtin_ia32_top4buud(W, A, B) __builtin_ia32_top4buud (1, A, B) +#define __builtin_ia32_top4mxbf8ps(W, A, B, C) __builtin_ia32_top4mxbf8ps (1, A, B, 9) +#define __builtin_ia32_top4mxbhf8ps(W, A, B, C) __builtin_ia32_top4mxbhf8ps (1, A, B, 9) +#define __builtin_ia32_top4mxhbf8ps(W, A, B, C) __builtin_ia32_top4mxhbf8ps (1, A, B, 9) +#define __builtin_ia32_top4mxhf8ps(W, A, B, C) __builtin_ia32_top4mxhf8ps (1, A, B, 9) +#define __builtin_ia32_top4mxbssps(W, A, B, C) __builtin_ia32_top4mxbssps (1, A, B, 9) #endif #include <wmmintrin.h> diff --git a/gcc/testsuite/gcc.target/i386/fp8-helper.h b/gcc/testsuite/gcc.target/i386/fp8-helper.h index d3f526f651c..f113ec76f4a 100644 --- a/gcc/testsuite/gcc.target/i386/fp8-helper.h +++ b/gcc/testsuite/gcc.target/i386/fp8-helper.h @@ -210,14 +210,19 @@ bf8_isInf(unsigned char value) static int64_t shift_fp8_to_int64 (unsigned char x, int bf8, int *valueState) { - *valueState = (x >> 7) ? -1 : 1; - if (fp8_isNan (x, bf8)) - *valueState = -2; - if (bf8) - if (bf8_isInf(x) == 1) - *valueState = 3; - else if (bf8_isInf(x) == -1) - *valueState = -3; + if (valueState != NULL) + { + *valueState = (x >> 7) ? -1 : 1; + if (fp8_isNan (x, bf8)) + *valueState = -2; + if (bf8) + { + if (bf8_isInf(x) == 1) + *valueState = 3; + else if (bf8_isInf(x) == -1) + *valueState = -3; + } + } unsigned short sign = (x & 0x80) >> 7; unsigned short exp = bf8 ? (x & 0x7c) >> 2 : (x & 0x78) >> 3; diff --git a/gcc/testsuite/gcc.target/i386/sse-13.c b/gcc/testsuite/gcc.target/i386/sse-13.c index 67ee7f0088d..3284fcd4bd4 100644 --- a/gcc/testsuite/gcc.target/i386/sse-13.c +++ b/gcc/testsuite/gcc.target/i386/sse-13.c @@ -936,6 +936,16 @@ #define __builtin_ia32_tilemovrowextract(A, B) __builtin_ia32_tilemovrowextract (1, B) #define __builtin_ia32_tilemovrowinsert(A, B, C) __builtin_ia32_tilemovrowinsert (1, B, C) #define __builtin_ia32_tilemovcolinsert(A, B, C) __builtin_ia32_tilemovcolinsert (1, B, C) +#define __builtin_ia32_top2bf16ps(W, A, B) __builtin_ia32_top2bf16ps (1, A, B) +#define __builtin_ia32_top4bssd(W, A, B) __builtin_ia32_top4bssd (1, A, B) +#define __builtin_ia32_top4bsud(W, A, B) __builtin_ia32_top4bsud (1, A, B) +#define __builtin_ia32_top4busd(W, A, B) __builtin_ia32_top4busd (1, A, B) +#define __builtin_ia32_top4buud(W, A, B) __builtin_ia32_top4buud (1, A, B) +#define __builtin_ia32_top4mxbf8ps(W, A, B, C) __builtin_ia32_top4mxbf8ps (1, A, B, 9) +#define __builtin_ia32_top4mxbhf8ps(W, A, B, C) __builtin_ia32_top4mxbhf8ps (1, A, B, 9) +#define __builtin_ia32_top4mxhbf8ps(W, A, B, C) __builtin_ia32_top4mxhbf8ps (1, A, B, 9) +#define __builtin_ia32_top4mxhf8ps(W, A, B, C) __builtin_ia32_top4mxhf8ps (1, A, B, 9) +#define __builtin_ia32_top4mxbssps(W, A, B, C) __builtin_ia32_top4mxbssps (1, A, B, 9) #endif #include <x86intrin.h> diff --git a/gcc/testsuite/gcc.target/i386/sse-14.c b/gcc/testsuite/gcc.target/i386/sse-14.c index c0b9fb0a796..1e9ad6bdcf9 100644 --- a/gcc/testsuite/gcc.target/i386/sse-14.c +++ b/gcc/testsuite/gcc.target/i386/sse-14.c @@ -65,6 +65,10 @@ void _CONCAT(_,func) (op1_type A, op2_type B, int const I, int const L) \ { func (A, B, imm1, imm2); } +#define test_2vxt(func, imm1, op1_type, op2_type, imm2) \ + void _CONCAT(_,func) (int const I, op1_type A, op2_type B, int const L) \ + { func (imm1, A, B, imm2); } + #define test_3(func, type, op1_type, op2_type, op3_type, imm) \ type _CONCAT(_,func) (op1_type A, op2_type B, \ op3_type C, int const I) \ @@ -1227,4 +1231,14 @@ test_1t (_tile_cvtrowl_ps_ph, __m512h, 1, int) test_1t (_tile_extractrow, __m512i, 1, int) test_2vt (_tile_insertrow, 1, __m512i, int) test_2vt (_tile_insertcol, 1, __m512i, int) +test_2vt (_tile_op2bf16_ps, 1, __m512bh, __m512bh) +test_2vt (_tile_op4bssd_epi32, 1, __m512i, __m512i) +test_2vt (_tile_op4bsud_epi32, 1, __m512i, __m512i) +test_2vt (_tile_op4busd_epi32, 1, __m512i, __m512i) +test_2vt (_tile_op4buud_epi32, 1, __m512i, __m512i) +test_2vxt (_tile_op4mxbf8_ps, 1, __m512i, __m512i, 9) +test_2vxt (_tile_op4mxbhf8_ps, 1, __m512i, __m512i, 9) +test_2vxt (_tile_op4mxhbf8_ps, 1, __m512i, __m512i, 9) +test_2vxt (_tile_op4mxhf8_ps, 1, __m512i, __m512i, 9) +test_2vxt (_tile_op4mxbss_ps, 1, __m512i, __m512i, 9) #endif diff --git a/gcc/testsuite/gcc.target/i386/sse-22.c b/gcc/testsuite/gcc.target/i386/sse-22.c index fa5e42fc808..9e50994946c 100644 --- a/gcc/testsuite/gcc.target/i386/sse-22.c +++ b/gcc/testsuite/gcc.target/i386/sse-22.c @@ -58,6 +58,10 @@ type _CONCAT(_,func) (op1_type A, op2_type B, int const I, int const L) \ { return func (A, B, imm1, imm2); } +#define test_2vxt(func, imm1, op1_type, op2_type, imm2) \ + void _CONCAT(_,func) (int const I, op1_type A, op2_type B, int const L) \ + { func (imm1, A, B, imm2); } + #define test_2y(func, type, op1_type, op2_type, imm1, imm2, imm3) \ type _CONCAT(_,func) (op1_type A, op2_type B, int const I, int const L,\ int const R) \ @@ -1268,4 +1272,14 @@ test_1t (_tile_cvtrowl_ps_ph, __m512h, 1, int) test_1t (_tile_extractrow, __m512i, 1, int) test_2vt (_tile_insertrow, 1, __m512i, int) test_2vt (_tile_insertcol, 1, __m512i, int) +test_2vt (_tile_op2bf16_ps, 1, __m512bh, __m512bh) +test_2vt (_tile_op4bssd_epi32, 1, __m512i, __m512i) +test_2vt (_tile_op4bsud_epi32, 1, __m512i, __m512i) +test_2vt (_tile_op4busd_epi32, 1, __m512i, __m512i) +test_2vt (_tile_op4buud_epi32, 1, __m512i, __m512i) +test_2vxt (_tile_op4mxbf8_ps, 1, __m512i, __m512i, 9) +test_2vxt (_tile_op4mxbhf8_ps, 1, __m512i, __m512i, 9) +test_2vxt (_tile_op4mxhbf8_ps, 1, __m512i, __m512i, 9) +test_2vxt (_tile_op4mxhf8_ps, 1, __m512i, __m512i, 9) +test_2vxt (_tile_op4mxbss_ps, 1, __m512i, __m512i, 9) #endif diff --git a/gcc/testsuite/gcc.target/i386/sse-23.c b/gcc/testsuite/gcc.target/i386/sse-23.c index 741984fded1..c6cf53b5ed0 100644 --- a/gcc/testsuite/gcc.target/i386/sse-23.c +++ b/gcc/testsuite/gcc.target/i386/sse-23.c @@ -911,6 +911,16 @@ #define __builtin_ia32_tilemovrowextract(A, B) __builtin_ia32_tilemovrowextract (1, B) #define __builtin_ia32_tilemovrowinsert(A, B, C) __builtin_ia32_tilemovrowinsert (1, B, C) #define __builtin_ia32_tilemovcolinsert(A, B, C) __builtin_ia32_tilemovcolinsert (1, B, C) +#define __builtin_ia32_top2bf16ps(W, A, B) __builtin_ia32_top2bf16ps (1, A, B) +#define __builtin_ia32_top4bssd(W, A, B) __builtin_ia32_top4bssd (1, A, B) +#define __builtin_ia32_top4bsud(W, A, B) __builtin_ia32_top4bsud (1, A, B) +#define __builtin_ia32_top4busd(W, A, B) __builtin_ia32_top4busd (1, A, B) +#define __builtin_ia32_top4buud(W, A, B) __builtin_ia32_top4buud (1, A, B) +#define __builtin_ia32_top4mxbf8ps(W, A, B, C) __builtin_ia32_top4mxbf8ps (1, A, B, 9) +#define __builtin_ia32_top4mxbhf8ps(W, A, B, C) __builtin_ia32_top4mxbhf8ps (1, A, B, 9) +#define __builtin_ia32_top4mxhbf8ps(W, A, B, C) __builtin_ia32_top4mxhbf8ps (1, A, B, 9) +#define __builtin_ia32_top4mxhf8ps(W, A, B, C) __builtin_ia32_top4mxhf8ps (1, A, B, 9) +#define __builtin_ia32_top4mxbssps(W, A, B, C) __builtin_ia32_top4mxbssps (1, A, B, 9) #endif #pragma GCC target ("sse4a,3dnow,avx,avx2,fma4,xop,aes,pclmul,popcnt,abm,lzcnt,bmi,bmi2,tbm,lwp,fsgsbase,rdrnd,f16c,fma,rtm,rdseed,prfchw,adx,fxsr,xsaveopt,sha,xsavec,xsaves,clflushopt,clwb,mwaitx,clzero,pku,sgx,rdpid,gfni,vpclmulqdq,pconfig,wbnoinvd,enqcmd,avx512vp2intersect,serialize,tsxldtrk,amx-tile,amx-int8,amx-bf16,kl,widekl,avxvnni,avxifma,avxvnniint8,avxneconvert,cmpccxadd,amx-fp16,prefetchi,raoint,amx-complex,avxvnniint16,sm3,sha512,sm4,avx10.2,amx-avx512,amx-fp8,movrs,amx-movrs,avx10v2aux,acev1") -- 2.31.1