SVE2.2 (or in streaming mode, SME2.2) adds support for zeroing predication
for all variants of the following FP-to-integer conversion instructions:
- FCVTZU (Floating-point convert to unsigned integer, rounding toward zero
(predicated))
- FCVTZS (Floating-point convert to signed integer, rounding toward zero
(predicated))
To implement this change, this patch adds a new alternative to patterns
involving the SVE_COND_FCVTI iterator and accepting an independent value
as the merge operand. The new alternative has the new zeroing-predication
forms as the output string and is only enabled when sve2p2_or_sme2p2 is
true in the target architecture.
The new ASM tests only cover the "_z" versions of the intrinsics and as
such all have the "_z" suffix in their name, and are grouped by type of
the destination operand.
gcc/ChangeLog:
* config/aarch64/aarch64-sve.md
(*cond_<optab>_nontrunc<SVE_FULL_F:mode><SVE_FULL_HSDI:mode>_relaxed):
New alternative for zeroing predication. Add `arch` attribute
to every alternative.
(*cond_<optab>_nontrunc<SVE_PARTIAL_F:mode><SVE_HSDI:mode>_relaxed):
Likewise.
(*cond_<optab>_nontrunc<SVE_FULL_F:mode><SVE_FULL_HSDI:mode>_strict):
Likewise.
(*cond_<optab>_trunc<VNx2DF_ONLY:mode><VNx4SI_ONLY:mode>):
Likewise.
(*cond_<optab>_trunc<VNx2DF_ONLY:mode><VNx2SI_ONLY:mode>_relaxed):
Likewise.
gcc/testsuite/ChangeLog:
* gcc.target/aarch64/sve2/acle/asm/cvt_s16_z.c: New test.
* gcc.target/aarch64/sve2/acle/asm/cvt_s32_z.c: Likewise.
* gcc.target/aarch64/sve2/acle/asm/cvt_s64_z.c: Likewise.
* gcc.target/aarch64/sve2/acle/asm/cvt_u16_z.c: Likewise.
* gcc.target/aarch64/sve2/acle/asm/cvt_u32_z.c: Likewise.
* gcc.target/aarch64/sve2/acle/asm/cvt_u64_z.c: Likewise.
---
gcc/config/aarch64/aarch64-sve.md | 45 +++++++------
.../aarch64/sve2/acle/asm/cvt_s16_z.c | 29 ++++++++
.../aarch64/sve2/acle/asm/cvt_s32_z.c | 67 +++++++++++++++++++
.../aarch64/sve2/acle/asm/cvt_s64_z.c | 67 +++++++++++++++++++
.../aarch64/sve2/acle/asm/cvt_u16_z.c | 29 ++++++++
.../aarch64/sve2/acle/asm/cvt_u32_z.c | 67 +++++++++++++++++++
.../aarch64/sve2/acle/asm/cvt_u64_z.c | 67 +++++++++++++++++++
7 files changed, 351 insertions(+), 20 deletions(-)
create mode 100644 gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_s16_z.c
create mode 100644 gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_s32_z.c
create mode 100644 gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_s64_z.c
create mode 100644 gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_u16_z.c
create mode 100644 gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_u32_z.c
create mode 100644 gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_u64_z.c
diff --git a/gcc/config/aarch64/aarch64-sve.md
b/gcc/config/aarch64/aarch64-sve.md
index cb6c850d4c9..928f081ed73 100644
--- a/gcc/config/aarch64/aarch64-sve.md
+++ b/gcc/config/aarch64/aarch64-sve.md
@@ -10705,10 +10705,11 @@
(match_operand:SVE_FULL_HSDI 3 "aarch64_simd_reg_or_zero")]
UNSPEC_SEL))]
"TARGET_SVE && <SVE_FULL_HSDI:elem_bits> >= <SVE_FULL_F:elem_bits>"
- {@ [ cons: =0 , 1 , 2 , 3 ; attrs: movprfx ]
- [ &w , Upl , w , 0 ; * ]
fcvtz<su>\t%0.<SVE_FULL_HSDI:Vetype>, %1/m, %2.<SVE_FULL_F:Vetype>
- [ &w , Upl , w , Dz ; yes ]
movprfx\t%0.<SVE_FULL_HSDI:Vetype>, %1/z,
%2.<SVE_FULL_HSDI:Vetype>\;fcvtz<su>\t%0.<SVE_FULL_HSDI:Vetype>, %1/m,
%2.<SVE_FULL_F:Vetype>
- [ ?&w , Upl , w , w ; yes ] movprfx\t%0,
%3\;fcvtz<su>\t%0.<SVE_FULL_HSDI:Vetype>, %1/m, %2.<SVE_FULL_F:Vetype>
+ {@ [ cons: =0 , 1 , 2 , 3 ; attrs: movprfx, arch ]
+ [ &w , Upl , w , 0 ; * , * ]
fcvtz<su>\t%0.<SVE_FULL_HSDI:Vetype>, %1/m, %2.<SVE_FULL_F:Vetype>
+ [ &w , Upl , w , Dz ; * , sve2p2_or_sme2p2 ]
fcvtz<su>\t%0.<SVE_FULL_HSDI:Vetype>, %1/z, %2.<SVE_FULL_F:Vetype>
+ [ &w , Upl , w , Dz ; yes , * ]
movprfx\t%0.<SVE_FULL_HSDI:Vetype>, %1/z,
%2.<SVE_FULL_HSDI:Vetype>\;fcvtz<su>\t%0.<SVE_FULL_HSDI:Vetype>, %1/m,
%2.<SVE_FULL_F:Vetype>
+ [ ?&w , Upl , w , w ; yes , * ] movprfx\t%0,
%3\;fcvtz<su>\t%0.<SVE_FULL_HSDI:Vetype>, %1/m, %2.<SVE_FULL_F:Vetype>
}
"&& !rtx_equal_p (operands[1], operands[4])"
{
@@ -10731,10 +10732,11 @@
UNSPEC_SEL))]
"TARGET_SVE
&& (~(<SVE_HSDI:self_mask> | <SVE_HSDI:narrower_mask>) &
<SVE_PARTIAL_F:self_mask>) == 0"
- {@ [ cons: =0 , 1 , 2 , 3 ; attrs: movprfx ]
- [ &w , Upl , w , 0 ; * ]
fcvtz<su>\t%0.<SVE_HSDI:Vetype>, %1/m, %2.<SVE_PARTIAL_F:Vetype>
- [ &w , Upl , w , Dz ; yes ]
movprfx\t%0.<SVE_HSDI:Vetype>, %1/z,
%2.<SVE_HSDI:Vetype>\;fcvtz<su>\t%0.<SVE_HSDI:Vetype>, %1/m,
%2.<SVE_PARTIAL_F:Vetype>
- [ ?&w , Upl , w , w ; yes ] movprfx\t%0,
%3\;fcvtz<su>\t%0.<SVE_HSDI:Vetype>, %1/m, %2.<SVE_PARTIAL_F:Vetype>
+ {@ [ cons: =0 , 1 , 2 , 3 ; attrs: movprfx, arch ]
+ [ &w , Upl , w , 0 ; * , * ]
fcvtz<su>\t%0.<SVE_HSDI:Vetype>, %1/m, %2.<SVE_PARTIAL_F:Vetype>
+ [ &w , Upl , w , Dz ; * , sve2p2_or_sme2p2 ]
fcvtz<su>\t%0.<SVE_HSDI:Vetype>, %1/z, %2.<SVE_PARTIAL_F:Vetype>
+ [ &w , Upl , w , Dz ; yes , * ]
movprfx\t%0.<SVE_HSDI:Vetype>, %1/z,
%2.<SVE_HSDI:Vetype>\;fcvtz<su>\t%0.<SVE_HSDI:Vetype>, %1/m,
%2.<SVE_PARTIAL_F:Vetype>
+ [ ?&w , Upl , w , w ; yes , * ] movprfx\t%0,
%3\;fcvtz<su>\t%0.<SVE_HSDI:Vetype>, %1/m, %2.<SVE_PARTIAL_F:Vetype>
}
"&& !rtx_equal_p (operands[1], operands[4])"
{
@@ -10754,10 +10756,11 @@
(match_operand:SVE_FULL_HSDI 3 "aarch64_simd_reg_or_zero")]
UNSPEC_SEL))]
"TARGET_SVE && <SVE_FULL_HSDI:elem_bits> >= <SVE_FULL_F:elem_bits>"
- {@ [ cons: =0 , 1 , 2 , 3 ; attrs: movprfx ]
- [ &w , Upl , w , 0 ; * ]
fcvtz<su>\t%0.<SVE_FULL_HSDI:Vetype>, %1/m, %2.<SVE_FULL_F:Vetype>
- [ &w , Upl , w , Dz ; yes ]
movprfx\t%0.<SVE_FULL_HSDI:Vetype>, %1/z,
%2.<SVE_FULL_HSDI:Vetype>\;fcvtz<su>\t%0.<SVE_FULL_HSDI:Vetype>, %1/m,
%2.<SVE_FULL_F:Vetype>
- [ ?&w , Upl , w , w ; yes ] movprfx\t%0,
%3\;fcvtz<su>\t%0.<SVE_FULL_HSDI:Vetype>, %1/m, %2.<SVE_FULL_F:Vetype>
+ {@ [ cons: =0 , 1 , 2 , 3 ; attrs: movprfx, arch ]
+ [ &w , Upl , w , 0 ; * , * ]
fcvtz<su>\t%0.<SVE_FULL_HSDI:Vetype>, %1/m, %2.<SVE_FULL_F:Vetype>
+ [ &w , Upl , w , Dz ; * , sve2p2_or_sme2p2 ]
fcvtz<su>\t%0.<SVE_FULL_HSDI:Vetype>, %1/z, %2.<SVE_FULL_F:Vetype>
+ [ &w , Upl , w , Dz ; yes , * ]
movprfx\t%0.<SVE_FULL_HSDI:Vetype>, %1/z,
%2.<SVE_FULL_HSDI:Vetype>\;fcvtz<su>\t%0.<SVE_FULL_HSDI:Vetype>, %1/m,
%2.<SVE_FULL_F:Vetype>
+ [ ?&w , Upl , w , w ; yes , * ] movprfx\t%0,
%3\;fcvtz<su>\t%0.<SVE_FULL_HSDI:Vetype>, %1/m, %2.<SVE_FULL_F:Vetype>
}
[(set_attr "sve_type" "sve_fp_cvt")]
)
@@ -10789,10 +10792,11 @@
(match_operand:VNx4SI_ONLY 3 "aarch64_simd_reg_or_zero")]
UNSPEC_SEL))]
"TARGET_SVE"
- {@ [ cons: =0 , 1 , 2 , 3 ; attrs: movprfx ]
- [ &w , Upl , w , 0 ; * ]
fcvtz<su>\t%0.<VNx4SI_ONLY:Vetype>, %1/m, %2.<VNx2DF_ONLY:Vetype>
- [ &w , Upl , w , Dz ; yes ]
movprfx\t%0.<VNx2DF_ONLY:Vetype>, %1/z,
%2.<VNx2DF_ONLY:Vetype>\;fcvtz<su>\t%0.<VNx4SI_ONLY:Vetype>, %1/m,
%2.<VNx2DF_ONLY:Vetype>
- [ ?&w , Upl , w , w ; yes ] movprfx\t%0,
%3\;fcvtz<su>\t%0.<VNx4SI_ONLY:Vetype>, %1/m, %2.<VNx2DF_ONLY:Vetype>
+ {@ [ cons: =0 , 1 , 2 , 3 ; attrs: movprfx, arch ]
+ [ &w , Upl , w , 0 ; * , * ]
fcvtz<su>\t%0.<VNx4SI_ONLY:Vetype>, %1/m, %2.<VNx2DF_ONLY:Vetype>
+ [ &w , Upl , w , Dz ; * , sve2p2_or_sme2p2 ]
fcvtz<su>\t%0.<VNx4SI_ONLY:Vetype>, %1/z, %2.<VNx2DF_ONLY:Vetype>
+ [ &w , Upl , w , Dz ; yes , * ]
movprfx\t%0.<VNx2DF_ONLY:Vetype>, %1/z,
%2.<VNx2DF_ONLY:Vetype>\;fcvtz<su>\t%0.<VNx4SI_ONLY:Vetype>, %1/m,
%2.<VNx2DF_ONLY:Vetype>
+ [ ?&w , Upl , w , w ; yes , * ] movprfx\t%0,
%3\;fcvtz<su>\t%0.<VNx4SI_ONLY:Vetype>, %1/m, %2.<VNx2DF_ONLY:Vetype>
}
[(set_attr "sve_type" "sve_fp_cvt")]
)
@@ -10809,10 +10813,11 @@
(match_operand:VNx2SI_ONLY 3 "aarch64_simd_reg_or_zero")]
UNSPEC_SEL))]
"TARGET_SVE"
- {@ [ cons: =0 , 1 , 2 , 3 ; attrs: movprfx ]
- [ &w , Upl , w , 0 ; * ]
fcvtz<su>\t%0.<VNx2SI_ONLY:Vetype>, %1/m, %2.<VNx2DF_ONLY:Vetype>
- [ &w , Upl , w , Dz ; yes ]
movprfx\t%0.<VNx2DF_ONLY:Vetype>, %1/z,
%2.<VNx2DF_ONLY:Vetype>\;fcvtz<su>\t%0.<VNx2SI_ONLY:Vetype>, %1/m,
%2.<VNx2DF_ONLY:Vetype>
- [ ?&w , Upl , w , w ; yes ] movprfx\t%0,
%3\;fcvtz<su>\t%0.<VNx2SI_ONLY:Vetype>, %1/m, %2.<VNx2DF_ONLY:Vetype>
+ {@ [ cons: =0 , 1 , 2 , 3 ; attrs: movprfx, arch ]
+ [ &w , Upl , w , 0 ; * , * ]
fcvtz<su>\t%0.<VNx2SI_ONLY:Vetype>, %1/m, %2.<VNx2DF_ONLY:Vetype>
+ [ &w , Upl , w , Dz ; * , sve2p2_or_sme2p2 ]
fcvtz<su>\t%0.<VNx2SI_ONLY:Vetype>, %1/z, %2.<VNx2DF_ONLY:Vetype>
+ [ &w , Upl , w , Dz ; yes , * ]
movprfx\t%0.<VNx2DF_ONLY:Vetype>, %1/z,
%2.<VNx2DF_ONLY:Vetype>\;fcvtz<su>\t%0.<VNx2SI_ONLY:Vetype>, %1/m,
%2.<VNx2DF_ONLY:Vetype>
+ [ ?&w , Upl , w , w ; yes , * ] movprfx\t%0,
%3\;fcvtz<su>\t%0.<VNx2SI_ONLY:Vetype>, %1/m, %2.<VNx2DF_ONLY:Vetype>
}
"&& !rtx_equal_p (operands[1], operands[4])"
{
diff --git a/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_s16_z.c
b/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_s16_z.c
new file mode 100644
index 00000000000..2fd898cbb4a
--- /dev/null
+++ b/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_s16_z.c
@@ -0,0 +1,29 @@
+/* { dg-do assemble { target aarch64_asm_sve2p2_ok } } */
+/* { dg-do compile { target { ! aarch64_asm_sve2p2_ok } } } */
+/* { dg-final { check-function-bodies "**" "" "-DCHECK_ASM" } } */
+
+#include "test_sve_acle.h"
+
+#pragma GCC target "+sve2p2"
+#ifdef STREAMING_COMPATIBLE
+#pragma GCC target "+sme2p2"
+#endif
+
+/*
+** cvt_s16_f16_z_tied1:
+** mov (z[0-9]+)\.d, z0\.d
+** fcvtzs z0\.h, p0/z, \1\.h
+** ret
+*/
+TEST_DUAL_Z_REV (cvt_s16_f16_z_tied1, svint16_t, svfloat16_t,
+ z0_res = svcvt_s16_f16_z (p0, z0),
+ z0_res = svcvt_s16_z (p0, z0))
+
+/*
+** cvt_s16_f16_z_untied:
+** fcvtzs z0\.h, p0/z, z4\.h
+** ret
+*/
+TEST_DUAL_Z (cvt_s16_f16_z_untied, svint16_t, svfloat16_t,
+ z0 = svcvt_s16_f16_z (p0, z4),
+ z0 = svcvt_s16_z (p0, z4))
diff --git a/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_s32_z.c
b/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_s32_z.c
new file mode 100644
index 00000000000..c123adeef00
--- /dev/null
+++ b/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_s32_z.c
@@ -0,0 +1,67 @@
+/* { dg-do assemble { target aarch64_asm_sve2p2_ok } } */
+/* { dg-do compile { target { ! aarch64_asm_sve2p2_ok } } } */
+/* { dg-final { check-function-bodies "**" "" "-DCHECK_ASM" } } */
+
+#include "test_sve_acle.h"
+
+#pragma GCC target "+sve2p2"
+#ifdef STREAMING_COMPATIBLE
+#pragma GCC target "+sme2p2"
+#endif
+
+/*
+** cvt_s32_f16_z_tied1:
+** mov (z[0-9]+)\.d, z0\.d
+** fcvtzs z0\.s, p0/z, \1\.h
+** ret
+*/
+TEST_DUAL_Z_REV (cvt_s32_f16_z_tied1, svint32_t, svfloat16_t,
+ z0_res = svcvt_s32_f16_z (p0, z0),
+ z0_res = svcvt_s32_z (p0, z0))
+
+/*
+** cvt_s32_f16_z_untied:
+** fcvtzs z0\.s, p0/z, z4\.h
+** ret
+*/
+TEST_DUAL_Z (cvt_s32_f16_z_untied, svint32_t, svfloat16_t,
+ z0 = svcvt_s32_f16_z (p0, z4),
+ z0 = svcvt_s32_z (p0, z4))
+
+/*
+** cvt_s32_f32_z_tied1:
+** mov (z[0-9]+)\.d, z0\.d
+** fcvtzs z0\.s, p0/z, \1\.s
+** ret
+*/
+TEST_DUAL_Z_REV (cvt_s32_f32_z_tied1, svint32_t, svfloat32_t,
+ z0_res = svcvt_s32_f32_z (p0, z0),
+ z0_res = svcvt_s32_z (p0, z0))
+
+/*
+** cvt_s32_f32_z_untied:
+** fcvtzs z0\.s, p0/z, z4\.s
+** ret
+*/
+TEST_DUAL_Z (cvt_s32_f32_z_untied, svint32_t, svfloat32_t,
+ z0 = svcvt_s32_f32_z (p0, z4),
+ z0 = svcvt_s32_z (p0, z4))
+
+/*
+** cvt_s32_f64_z_tied1:
+** mov (z[0-9]+\.d), z0\.d
+** fcvtzs z0\.s, p0/z, \1
+** ret
+*/
+TEST_DUAL_Z_REV (cvt_s32_f64_z_tied1, svint32_t, svfloat64_t,
+ z0_res = svcvt_s32_f64_z (p0, z0),
+ z0_res = svcvt_s32_z (p0, z0))
+
+/*
+** cvt_s32_f64_z_untied:
+** fcvtzs z0\.s, p0/z, z4\.d
+** ret
+*/
+TEST_DUAL_Z (cvt_s32_f64_z_untied, svint32_t, svfloat64_t,
+ z0 = svcvt_s32_f64_z (p0, z4),
+ z0 = svcvt_s32_z (p0, z4))
diff --git a/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_s64_z.c
b/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_s64_z.c
new file mode 100644
index 00000000000..65c13d90b26
--- /dev/null
+++ b/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_s64_z.c
@@ -0,0 +1,67 @@
+/* { dg-do assemble { target aarch64_asm_sve2p2_ok } } */
+/* { dg-do compile { target { ! aarch64_asm_sve2p2_ok } } } */
+/* { dg-final { check-function-bodies "**" "" "-DCHECK_ASM" } } */
+
+#include "test_sve_acle.h"
+
+#pragma GCC target "+sve2p2"
+#ifdef STREAMING_COMPATIBLE
+#pragma GCC target "+sme2p2"
+#endif
+
+/*
+** cvt_s64_f16_z_tied1:
+** mov (z[0-9]+)\.d, z0\.d
+** fcvtzs z0\.d, p0/z, \1\.h
+** ret
+*/
+TEST_DUAL_Z_REV (cvt_s64_f16_z_tied1, svint64_t, svfloat16_t,
+ z0_res = svcvt_s64_f16_z (p0, z0),
+ z0_res = svcvt_s64_z (p0, z0))
+
+/*
+** cvt_s64_f16_z_untied:
+** fcvtzs z0\.d, p0/z, z4\.h
+** ret
+*/
+TEST_DUAL_Z (cvt_s64_f16_z_untied, svint64_t, svfloat16_t,
+ z0 = svcvt_s64_f16_z (p0, z4),
+ z0 = svcvt_s64_z (p0, z4))
+
+/*
+** cvt_s64_f32_z_tied1:
+** mov (z[0-9]+)\.d, z0\.d
+** fcvtzs z0\.d, p0/z, \1\.s
+** ret
+*/
+TEST_DUAL_Z_REV (cvt_s64_f32_z_tied1, svint64_t, svfloat32_t,
+ z0_res = svcvt_s64_f32_z (p0, z0),
+ z0_res = svcvt_s64_z (p0, z0))
+
+/*
+** cvt_s64_f32_z_untied:
+** fcvtzs z0\.d, p0/z, z4\.s
+** ret
+*/
+TEST_DUAL_Z (cvt_s64_f32_z_untied, svint64_t, svfloat32_t,
+ z0 = svcvt_s64_f32_z (p0, z4),
+ z0 = svcvt_s64_z (p0, z4))
+
+/*
+** cvt_s64_f64_z_tied1:
+** mov (z[0-9]+\.d), z0\.d
+** fcvtzs z0\.d, p0/z, \1
+** ret
+*/
+TEST_DUAL_Z_REV (cvt_s64_f64_z_tied1, svint64_t, svfloat64_t,
+ z0_res = svcvt_s64_f64_z (p0, z0),
+ z0_res = svcvt_s64_z (p0, z0))
+
+/*
+** cvt_s64_f64_z_untied:
+** fcvtzs z0\.d, p0/z, z4\.d
+** ret
+*/
+TEST_DUAL_Z (cvt_s64_f64_z_untied, svint64_t, svfloat64_t,
+ z0 = svcvt_s64_f64_z (p0, z4),
+ z0 = svcvt_s64_z (p0, z4))
diff --git a/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_u16_z.c
b/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_u16_z.c
new file mode 100644
index 00000000000..aa625fd6f34
--- /dev/null
+++ b/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_u16_z.c
@@ -0,0 +1,29 @@
+/* { dg-do assemble { target aarch64_asm_sve2p2_ok } } */
+/* { dg-do compile { target { ! aarch64_asm_sve2p2_ok } } } */
+/* { dg-final { check-function-bodies "**" "" "-DCHECK_ASM" } } */
+
+#include "test_sve_acle.h"
+
+#pragma GCC target "+sve2p2"
+#ifdef STREAMING_COMPATIBLE
+#pragma GCC target "+sme2p2"
+#endif
+
+/*
+** cvt_u16_f16_z_tied1:
+** mov (z[0-9]+)\.d, z0\.d
+** fcvtzu z0\.h, p0/z, \1\.h
+** ret
+*/
+TEST_DUAL_Z_REV (cvt_u16_f16_z_tied1, svuint16_t, svfloat16_t,
+ z0_res = svcvt_u16_f16_z (p0, z0),
+ z0_res = svcvt_u16_z (p0, z0))
+
+/*
+** cvt_u16_f16_z_untied:
+** fcvtzu z0\.h, p0/z, z4\.h
+** ret
+*/
+TEST_DUAL_Z (cvt_u16_f16_z_untied, svuint16_t, svfloat16_t,
+ z0 = svcvt_u16_f16_z (p0, z4),
+ z0 = svcvt_u16_z (p0, z4))
diff --git a/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_u32_z.c
b/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_u32_z.c
new file mode 100644
index 00000000000..f2b84689760
--- /dev/null
+++ b/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_u32_z.c
@@ -0,0 +1,67 @@
+/* { dg-do assemble { target aarch64_asm_sve2p2_ok } } */
+/* { dg-do compile { target { ! aarch64_asm_sve2p2_ok } } } */
+/* { dg-final { check-function-bodies "**" "" "-DCHECK_ASM" } } */
+
+#include "test_sve_acle.h"
+
+#pragma GCC target "+sve2p2"
+#ifdef STREAMING_COMPATIBLE
+#pragma GCC target "+sme2p2"
+#endif
+
+/*
+** cvt_u32_f16_z_tied1:
+** mov (z[0-9]+)\.d, z0\.d
+** fcvtzu z0\.s, p0/z, \1\.h
+** ret
+*/
+TEST_DUAL_Z_REV (cvt_u32_f16_z_tied1, svuint32_t, svfloat16_t,
+ z0_res = svcvt_u32_f16_z (p0, z0),
+ z0_res = svcvt_u32_z (p0, z0))
+
+/*
+** cvt_u32_f16_z_untied:
+** fcvtzu z0\.s, p0/z, z4\.h
+** ret
+*/
+TEST_DUAL_Z (cvt_u32_f16_z_untied, svuint32_t, svfloat16_t,
+ z0 = svcvt_u32_f16_z (p0, z4),
+ z0 = svcvt_u32_z (p0, z4))
+
+/*
+** cvt_u32_f32_z_tied1:
+** mov (z[0-9]+)\.d, z0\.d
+** fcvtzu z0\.s, p0/z, \1\.s
+** ret
+*/
+TEST_DUAL_Z_REV (cvt_u32_f32_z_tied1, svuint32_t, svfloat32_t,
+ z0_res = svcvt_u32_f32_z (p0, z0),
+ z0_res = svcvt_u32_z (p0, z0))
+
+/*
+** cvt_u32_f32_z_untied:
+** fcvtzu z0\.s, p0/z, z4\.s
+** ret
+*/
+TEST_DUAL_Z (cvt_u32_f32_z_untied, svuint32_t, svfloat32_t,
+ z0 = svcvt_u32_f32_z (p0, z4),
+ z0 = svcvt_u32_z (p0, z4))
+
+/*
+** cvt_u32_f64_z_tied1:
+** mov (z[0-9]+\.d), z0\.d
+** fcvtzu z0\.s, p0/z, \1
+** ret
+*/
+TEST_DUAL_Z_REV (cvt_u32_f64_z_tied1, svuint32_t, svfloat64_t,
+ z0_res = svcvt_u32_f64_z (p0, z0),
+ z0_res = svcvt_u32_z (p0, z0))
+
+/*
+** cvt_u32_f64_z_untied:
+** fcvtzu z0\.s, p0/z, z4\.d
+** ret
+*/
+TEST_DUAL_Z (cvt_u32_f64_z_untied, svuint32_t, svfloat64_t,
+ z0 = svcvt_u32_f64_z (p0, z4),
+ z0 = svcvt_u32_z (p0, z4))
diff --git a/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_u64_z.c
b/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_u64_z.c
new file mode 100644
index 00000000000..5aa0e9bed89
--- /dev/null
+++ b/gcc/testsuite/gcc.target/aarch64/sve2/acle/asm/cvt_u64_z.c
@@ -0,0 +1,67 @@
+/* { dg-do assemble { target aarch64_asm_sve2p2_ok } } */
+/* { dg-do compile { target { ! aarch64_asm_sve2p2_ok } } } */
+/* { dg-final { check-function-bodies "**" "" "-DCHECK_ASM" } } */
+
+#include "test_sve_acle.h"
+
+#pragma GCC target "+sve2p2"
+#ifdef STREAMING_COMPATIBLE
+#pragma GCC target "+sme2p2"
+#endif
+
+/*
+** cvt_u64_f16_z_tied1:
+** mov (z[0-9]+)\.d, z0\.d
+** fcvtzu z0\.d, p0/z, \1\.h
+** ret
+*/
+TEST_DUAL_Z_REV (cvt_u64_f16_z_tied1, svuint64_t, svfloat16_t,
+ z0_res = svcvt_u64_f16_z (p0, z0),
+ z0_res = svcvt_u64_z (p0, z0))
+
+/*
+** cvt_u64_f16_z_untied:
+** fcvtzu z0\.d, p0/z, z4\.h
+** ret
+*/
+TEST_DUAL_Z (cvt_u64_f16_z_untied, svuint64_t, svfloat16_t,
+ z0 = svcvt_u64_f16_z (p0, z4),
+ z0 = svcvt_u64_z (p0, z4))
+
+/*
+** cvt_u64_f32_z_tied1:
+** mov (z[0-9]+)\.d, z0\.d
+** fcvtzu z0\.d, p0/z, \1\.s
+** ret
+*/
+TEST_DUAL_Z_REV (cvt_u64_f32_z_tied1, svuint64_t, svfloat32_t,
+ z0_res = svcvt_u64_f32_z (p0, z0),
+ z0_res = svcvt_u64_z (p0, z0))
+
+/*
+** cvt_u64_f32_z_untied:
+** fcvtzu z0\.d, p0/z, z4\.s
+** ret
+*/
+TEST_DUAL_Z (cvt_u64_f32_z_untied, svuint64_t, svfloat32_t,
+ z0 = svcvt_u64_f32_z (p0, z4),
+ z0 = svcvt_u64_z (p0, z4))
+
+/*
+** cvt_u64_f64_z_tied1:
+** mov (z[0-9]+\.d), z0\.d
+** fcvtzu z0\.d, p0/z, \1
+** ret
+*/
+TEST_DUAL_Z_REV (cvt_u64_f64_z_tied1, svuint64_t, svfloat64_t,
+ z0_res = svcvt_u64_f64_z (p0, z0),
+ z0_res = svcvt_u64_z (p0, z0))
+
+/*
+** cvt_u64_f64_z_untied:
+** fcvtzu z0\.d, p0/z, z4\.d
+** ret
+*/
+TEST_DUAL_Z (cvt_u64_f64_z_untied, svuint64_t, svfloat64_t,
+ z0 = svcvt_u64_f64_z (p0, z4),
+ z0 = svcvt_u64_z (p0, z4))
--
2.43.0
