Re: [PATCH] vect: Use signed offsets for negative step (PR121393)

[Andrew Stubbs]

On 02/10/2025 10:58, Richard Biener wrote:
On Fri, Sep 26, 2025 at 1:27 PM Andrew Stubbs <a...@baylibre.com> wrote:

When testcases (such as libgomp.c-c++-common/for-11.c) use unsigned types for
loop counters then the vectorizer will also use unsigned types for the offsets
passed to gather/scatter instructions.  This is a problem if the loop step is
negative and the offset type is narrower than a pointer.

Hmm, but why did we chose an unsigned type in the first place then?  So,
isn't this an issue in vect_truncate_gather_scatter_offset or
vect_use_strided_gather_scatters_p?

Here's the reproducer from PR121393:

float a[100];

int f()
{
  unsigned int i;                  <----- unsigned comes from here

#pragma omp target simd
  for (i=100; i != 0; i-=2)
    a[i] = a[i] + 5.5;
}

int main()
{
  f();
  return 0;
}


The patch is not needed if "i" is either signed or long long.

I have not observed issues in the other places, but I don't claim a full 
understanding of this code.

Andrew


This commit ensures that negative offsets are labelled as signed.

gcc/ChangeLog:

         PR target/121393
         * tree-vect-stmts.cc (vect_get_strided_load_store_ops): Use signed
         offsets for negative steps.
---

This fixes a problem observed on amdgcn, but there doesn't seem to be
anything about it that is target-specific.

OK for mainline?

Andrew


  gcc/tree-vect-stmts.cc | 8 ++++++++
  1 file changed, 8 insertions(+)

diff --git a/gcc/tree-vect-stmts.cc b/gcc/tree-vect-stmts.cc
index cfc4f323a22..254fe6e6c0f 100644
--- a/gcc/tree-vect-stmts.cc
+++ b/gcc/tree-vect-stmts.cc
@@ -2969,6 +2969,14 @@ vect_get_strided_load_store_ops (stmt_vec_info 
stmt_info, slp_tree node,
       type of the vector instead.  */
    tree offset_type = TREE_TYPE (offset_vectype);

+  if (TREE_CODE (DR_STEP (dr)) == INTEGER_CST
+      && compare_step_with_zero (loop_vinfo, stmt_info) < 0
+      && TYPE_SIGN (offset_type) == UNSIGNED)
+    {
+      offset_type = signed_type_for (offset_type);
+      offset_vectype = signed_type_for (offset_vectype);
+    }
+
    /* Calculate X = DR_STEP / SCALE and convert it to the appropriate type.  */
    tree step = size_binop (EXACT_DIV_EXPR, unshare_expr (DR_STEP (dr)),
                           ssize_int (SLP_TREE_GS_SCALE (node)));
--
2.51.0