[Bug fortran/102330] New: [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932

[Bug fortran/102330] New: [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932

[gscfq@t-online.de]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

            Bug ID: 102330
           Summary: [12 Regression] ICE in expand_gimple_stmt_1, at
                    cfgexpand.c:3932
           Product: gcc
           Version: 12.0
            Status: UNCONFIRMED
          Severity: normal
          Priority: P3
         Component: fortran
          Assignee: unassigned at gcc dot gnu.org
          Reporter: gscfq@t-online.de
  Target Milestone: ---

Started between 20210530 and 20210606 :


$ cat z1.f90
program p
  !$omp master taskloop simd
  do i = 1, 8
  end do
  !$acc parallel loop
  do i = 1, 8
  end do
end


$ gfortran-12-20210905 -c z1.f90 -fopenmp -fopenacc
during RTL pass: expand
z1.f90:5:21:

    5 |   !$acc parallel loop
      |                     ^
internal compiler error: in expand_gimple_stmt_1, at cfgexpand.c:3932
0x8dd6db expand_gimple_stmt_1
        ../../gcc/cfgexpand.c:3931
0x8dd6db expand_gimple_stmt
        ../../gcc/cfgexpand.c:4040
0x8e2bd7 expand_gimple_basic_block
        ../../gcc/cfgexpand.c:6082
0x8e553e execute
        ../../gcc/cfgexpand.c:6808

---

z1.f90:8:3:

    8 | end
      |   ^
Error: non-register as LHS of binary operation
# .MEM_25 = VDEF <.MEM_3>
i = 1 + .offset.9_2;
z1.f90:8:3: Error: non-register as LHS of binary operation
# .MEM_28 = VDEF <.MEM_4>
i = 1 + 8;
during GIMPLE pass: ssa
z1.f90:8:3: internal compiler error: verify_gimple failed
0xf05dd4 verify_gimple_in_cfg(function*, bool)
        ../../gcc/tree-cfg.c:5531
0xdb304e execute_function_todo
        ../../gcc/passes.c:2042
0xdb3ef2 execute_todo
        ../../gcc/passes.c:2096

[Bug fortran/102330] [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932

[rguenth at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

Richard Biener <rguenth at gcc dot gnu.org> changed:

           What    |Removed                     |Added
----------------------------------------------------------------------------
     Ever confirmed|0                           |1
   Last reconfirmed|                            |2021-09-15
             Status|UNCONFIRMED                 |NEW

--- Comment #1 from Richard Biener <rguenth at gcc dot gnu.org> ---
Confirmed.  omp expansion seems to introudce this non-gimple code and likely
also makes 'i' not a register (for OACC):

  .data_dep.5D.4044 = .UNIQUE (OACC_PRIVATE, .data_dep.5D.4044, -1, &iD.4045);

but

  iD.4045 = 1 + .offset.9D.4049;

and

  iD.4045 = 1 + 8;

(which is also unfolded)

[Bug fortran/102330] [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932

[rguenth at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

Richard Biener <rguenth at gcc dot gnu.org> changed:

           What    |Removed                     |Added
----------------------------------------------------------------------------
   Target Milestone|---                         |12.0

[Bug fortran/102330] [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932

[rguenth at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

Richard Biener <rguenth at gcc dot gnu.org> changed:

           What    |Removed                     |Added
----------------------------------------------------------------------------
                 CC|                            |jakub at gcc dot gnu.org

--- Comment #2 from Richard Biener <rguenth at gcc dot gnu.org> ---
Btw, GCC 11 rejects the testcase with

Error: Unexpected junk after $OMP MASTER statement at (1)

[Bug fortran/102330] [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932 since r12-1139-gf6bf436d9ab907d0

[marxin at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

Martin Liška <marxin at gcc dot gnu.org> changed:

           What    |Removed                     |Added
----------------------------------------------------------------------------
            Summary|[12 Regression] ICE in      |[12 Regression] ICE in
                   |expand_gimple_stmt_1, at    |expand_gimple_stmt_1, at
                   |cfgexpand.c:3932            |cfgexpand.c:3932 since
                   |                            |r12-1139-gf6bf436d9ab907d0
                 CC|                            |burnus at gcc dot gnu.org,
                   |                            |marxin at gcc dot gnu.org

--- Comment #3 from Martin Liška <marxin at gcc dot gnu.org> ---
Started with r12-1139-gf6bf436d9ab907d0.

[Bug fortran/102330] [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932 since r12-1139-gf6bf436d9ab907d0

[jakub at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

--- Comment #4 from Jakub Jelinek <jakub at gcc dot gnu.org> ---
Just
program p
  !$omp taskloop lastprivate(i)
  do i = 1, 8
  end do
  !$acc parallel loop
  do i = 1, 8
  end do
end
is enough to reproduce it.

[Bug fortran/102330] [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932 since r12-1139-gf6bf436d9ab907d0

[tschwinge at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

Thomas Schwinge <tschwinge at gcc dot gnu.org> changed:

           What    |Removed                     |Added
----------------------------------------------------------------------------
                 CC|                            |jules at gcc dot gnu.org,
                   |                            |tschwinge at gcc dot gnu.org

--- Comment #5 from Thomas Schwinge <tschwinge at gcc dot gnu.org> ---
(In reply to Jakub Jelinek from comment #4)
> Just
> program p
>   !$omp taskloop lastprivate(i)
>   do i = 1, 8
>   end do
>   !$acc parallel loop
>   do i = 1, 8
>   end do
> end
> is enough to reproduce it.

Confirmed.

With '-fopt-info-omp-all --param=openacc-privatization=noisy' we're told:
'note: variable ‘i’ in ‘private’ clause is candidate for adjusting OpenACC
privatization level', and...

(In reply to Richard Biener from comment #1)
> Confirmed.  omp expansion seems to introudce this non-gimple code and likely
> also makes 'i' not a register (for OACC):
> 
>   .data_dep.5D.4044 = .UNIQUE (OACC_PRIVATE, .data_dep.5D.4044, -1,
> &iD.4045);
> 
> but
> 
>   iD.4045 = 1 + .offset.9D.4049;
> 
> and
> 
>   iD.4045 = 1 + 8;
> 
> (which is also unfolded)

... indeed the ICE goes away if I force-disable 'OACC_PRIVATE' commit
29a2f51806c5b30e17a8d0e9ba7915a3c53c34ff "openacc: Add support for gang local
storage allocation in shared memory [PR90115]" as follows:

    --- gcc/omp-low.c
    +++ gcc/omp-low.c
    @@ -10494,6 +10494,8 @@ static bool
     oacc_privatization_candidate_p (const location_t loc, const tree c,
                                    const tree decl)
     {
    +  return false;

Hmm.  Julian CCed.

[Bug fortran/102330] [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932 since r12-980-g29a2f51806c

[burnus at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

Tobias Burnus <burnus at gcc dot gnu.org> changed:

           What    |Removed                     |Added
----------------------------------------------------------------------------
                 CC|                            |cltang at gcc dot gnu.org
           Keywords|                            |ice-on-valid-code
            Summary|[12 Regression] ICE in      |[12 Regression] ICE in
                   |expand_gimple_stmt_1, at    |expand_gimple_stmt_1, at
                   |cfgexpand.c:3932 since      |cfgexpand.c:3932 since
                   |r12-1139-gf6bf436d9ab907d0  |r12-980-g29a2f51806c

--- Comment #6 from Tobias Burnus <burnus at gcc dot gnu.org> ---
(In reply to Jakub Jelinek from comment #4)
> Just
...
> is enough to reproduce it.

This uses older features. Bisecting points now to:

  r12-980-g29a2f51806c
  openacc: Add support for gang local storage allocation in shared memory
[PR90115]

[Bug fortran/102330] [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932 since r12-980-g29a2f51806c

[jakub at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

--- Comment #7 from Jakub Jelinek <jakub at gcc dot gnu.org> ---
program p
  i = 0
  !$omp task shared(i)
  i = 1
  !$omp end task
  !$omp taskwait
  !$acc parallel loop
  do i = 1, 8
  end do
end

also ICEs, it just needs the iterator that is marked addressable during
omp-low.c which has code to regimplify statements that access such variables,
but it seems the problematic stmts are emitted only during omp-expand.c.  The
various expand_omp_* loop expansions have DECL_P (...) && TREE_ADDRESSABLE
(...) conditions in various spots, but perhaps OpenACC loop expansion doesn't.

[Bug fortran/102330] [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932 since r12-980-g29a2f51806c

[rguenth at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

Richard Biener <rguenth at gcc dot gnu.org> changed:

           What    |Removed                     |Added
----------------------------------------------------------------------------
           Priority|P3                          |P1

[Bug fortran/102330] [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932 since r12-980-g29a2f51806c

[tschwinge at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

Thomas Schwinge <tschwinge at gcc dot gnu.org> changed:

           What    |Removed                     |Added
----------------------------------------------------------------------------
             Status|NEW                         |ASSIGNED
           Assignee|unassigned at gcc dot gnu.org      |tschwinge at gcc dot gnu.org

[Bug fortran/102330] [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932 since r12-980-g29a2f51806c

[tschwinge at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

Thomas Schwinge <tschwinge at gcc dot gnu.org> changed:

           What    |Removed                     |Added
----------------------------------------------------------------------------
                 CC|                            |rguenth at gcc dot gnu.org

--- Comment #8 from Thomas Schwinge <tschwinge at gcc dot gnu.org> ---
Posting my current status here, in case anybody feels like verifying my
thoughts.  :-)

So I'm looking into Jakub's last, simplest reproducer:

(In reply to Jakub Jelinek from comment #7)
> program p
>   i = 0
>   !$omp task shared(i)
>   i = 1
>   !$omp end task
>   !$omp taskwait
>   !$acc parallel loop
>   do i = 1, 8
>   end do
> end
> 
> also ICEs

ACK.

> it just needs the iterator that is marked addressable during
> omp-low.c which has code to regimplify statements that access such
> variables, but it seems the problematic stmts are emitted only during
> omp-expand.c.

ACK, 'gcc/omp-expand.cc:expand_oacc_for'.

> The various expand_omp_* loop expansions have DECL_P (...) &&
> TREE_ADDRESSABLE (...) conditions in various spots, but perhaps OpenACC loop
> expansion doesn't.

(Thanks, that has help me with PR104132!)

However, that doesn't seem to be the problem here: we get reported
"non-register as LHS of binary operation", but we've got in
'gcc/omp-expand.cc:expand_oacc_for':

    7868          ass = gimple_build_assign (v, expr);
    (gdb) n
    7869          gsi_insert_before (&gsi, ass, GSI_SAME_STMT);
    (gdb) call debug_gimple_stmt(ass)
    i = 1 + .offset.8;
    (gdb) call debug_tree(v)
     <var_decl 0x7ffff7fc3e10 i
        type <integer_type 0x7ffff75555e8 integer(kind=4) public SI
            size <integer_cst 0x7ffff753ee40 constant 32>
            unit-size <integer_cst 0x7ffff753ee58 constant 4>
            align:32 warn_if_not_align:0 symtab:0 alias-set -1 canonical-type
0x7ffff75555e8 precision:32 min <integer_cst 0x7ffff753edf8 -2147483648> max
<integer_cst 0x7ffff753ee10 2147483647>
            pointer_to_this <pointer_type 0x7ffff755da80>>
        used SI
source-gcc/gcc/testsuite/gfortran.dg/goacc-gomp/pr102330-3.f90:9:3 size
<integer_cst 0x7ffff753ee40 32> unit-size <integer_cst 0x7ffff753ee58 4>
        align:32 warn_if_not_align:0 context <function_decl 0x7ffff7730d00 p>>
    (gdb) call is_gimple_lvalue(v)
    $4 = true
    (gdb) call is_gimple_reg(v)
    $5 = true
    (gdb) print v.base.addressable_flag 
    $6 = 0

So, 'i' is not 'TREE_ADDRESSABLE'.

Is the problem rather, that we earlier have done wrong, as Richard pointed out:

(In reply to Richard Biener from comment #1)
> Confirmed.  omp expansion seems to introudce this non-gimple code and likely also makes 'i' not a register (for OACC):
> 
>   .data_dep.5D.4044 = .UNIQUE (OACC_PRIVATE, .data_dep.5D.4044, -1, &iD.4045);

That's built in 'gcc/omp-low.cc:lower_oacc_private_marker'.  It's not marked
'TREE_ADDRESSABLE' (per debug log above), but the address of 'i' has been taken
(here).

> but
> 
>   iD.4045 = 1 + .offset.9D.4049;
> 
> and
> 
>   iD.4045 = 1 + 8;
> 
> (which is also unfolded)

Assuming 'TREE_ADDRESSABLE', that last item is resolved via my proposed
PR104132 change:

    [...]
    +gimple_simplified to D.4257 = 9;
    [...]
    -  iD.4257 = 1 + 8;
    +  D.4266 = 9;
    +  iD.4259 = D.4266;
    [...]

So if in 'gcc/omp-expand.cc:expand_oacc_for' I force 'TREE_ADDRESSABLE', then
the ICE goes away, and we get valid GIMPLE generated, per my understanding.

So, something like this maybe (untested)?  

    --- gcc/omp-low.cc
    +++ gcc/omp-low.cc
    @@ -11513,6 +11513,7 @@ lower_oacc_private_marker (omp_context *ctx)
            }
           gcc_checking_assert (decl);

    +      TREE_ADDRESSABLE (decl) = 1;
           tree addr = build_fold_addr_expr (decl);
           args.safe_push (addr);
         }

However: given that the '#pragma acc loop' has an implicit 'private(i)' -- why
does the 'gcc/omp-expand.cc:expand_oacc_for' even bother about the "outer 'i'"?
 The loop variable 'i' should be a completely separate, local, register-like
temporary.  Amy I misunderstanding something, or what's going wrong, elsewhere?

[Bug fortran/102330] [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932 since r12-980-g29a2f51806c

[jakub at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

--- Comment #9 from Jakub Jelinek <jakub at gcc dot gnu.org> ---
If you need to mark some var as addressable during omp lowering, then you need
to treat it similarly to the task shared case, so during scan phase of that
pass
do something like:
              /* Taking address of OUTER in lower_send_shared_vars
                 might need regimplification of everything that uses the
                 variable.  */
              if (!task_shared_vars)
                task_shared_vars = BITMAP_ALLOC (NULL);
              bitmap_set_bit (task_shared_vars, DECL_UID (outer));
              TREE_ADDRESSABLE (outer) = 1;
which will then make sure that the lowering phase of the pass will try to
regimplify all statements refering to such vars.  Note, this should be only
done if the var isn't already TREE_ADDRESSABLE.

We then do:

static tree
lower_omp_regimplify_p (tree *tp, int *walk_subtrees,
                        void *data)
{
  tree t = *tp;

...

  if (task_shared_vars
      && DECL_P (t)
      && bitmap_bit_p (task_shared_vars, DECL_UID (t)))
    return t;

Whether you can use the same bitmap or need to add another bitmap next to
task_shared_vars is something hard to guess without diving into it deeply.

[Bug fortran/102330] [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932 since r12-980-g29a2f51806c

[jakub at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

--- Comment #10 from Jakub Jelinek <jakub at gcc dot gnu.org> ---
Of course exceptions would be vars that certainly don't appear in the IL yet,
what I wrote about are vars that may appear there already.
Generally, vars should be marked as addressable before gimplification and
gimplification will then take the addressability into consideration, it is
wrong to mark vars as addressable in the middle of gimplification if they could
be already referenced earlier in the IL.  And similarly in omp-low.cc we
perform regimplifications and so stuff is fixable there too, but it needs to
know it should fix it up.

[Bug fortran/102330] [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932 since r12-980-g29a2f51806c

[tschwinge at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

--- Comment #11 from Thomas Schwinge <tschwinge at gcc dot gnu.org> ---
Jakub, thanks again for your comments!


I've written down the following while working through the issue.  Maybe you'd
like to verify, and maybe it'll useful for someone later (if only my future
self...).


(In reply to Thomas Schwinge from comment #8)
> (In reply to Richard Biener from comment #1)
> > Confirmed.  omp expansion seems to introudce this non-gimple code and likely also makes 'i' not a register (for OACC):
> > 
> >   .data_dep.5D.4044 = .UNIQUE (OACC_PRIVATE, .data_dep.5D.4044, -1, &iD.4045);
> 
> That's built in 'gcc/omp-low.cc:lower_oacc_private_marker'.  It's not marked
> 'TREE_ADDRESSABLE' (per debug log above), but the address of 'i' has been
> taken (here).

> So if in 'gcc/omp-expand.cc:expand_oacc_for' I force 'TREE_ADDRESSABLE',
> then the ICE goes away, and we get valid GIMPLE generated, per my
> understanding.
> 
> So, something like this maybe (untested)?  
> 
>     --- gcc/omp-low.cc
>     +++ gcc/omp-low.cc
>     @@ -11513,6 +11513,7 @@ lower_oacc_private_marker (omp_context *ctx)
>             }
>            gcc_checking_assert (decl);
>      
>     +      TREE_ADDRESSABLE (decl) = 1;
>            tree addr = build_fold_addr_expr (decl);
>            args.safe_push (addr);
>          }

That's bogus: per 'gcc/omp-low.cc:oacc_privatization_candidate_p', we should
only be getting here for 'TREE_ADDRESSABLE' DECLs.  Adding such an 'assert', it
triggers exactly for test cases like those we're discussing here.

However, before we look into the OpenACC privatization details, let's first
turn your test case into an OpenMP-only code, as follows:

(Based on Jakub Jelinek from comment #7)
>  program p
>    i = 0
>    !$omp task shared(i)
>    i = 1
>    !$omp end task
>    !$omp taskwait
> -  !$acc parallel loop
> +  !$omp parallel do
>    do i = 1, 8
> +     call f(i)
>    end do
> +  contains
> +
> +    subroutine f(v)
> +      integer :: v
> +      !integer, value :: v
> +    end
>  end

The 'call f(i)' serves as a convenient means to toggle whether or not 'i' in
the loop body needs to be addressable.

Using the 'integer :: v' variant, the 'call f(i)' is by-reference, and thus 'i'
is made "early 'TREE_ADDRESSABLE'".

Relevant part of '*.original' dump:

    [...]
    void p ()
    {
      [...]
      integer(kind=4)D.8 iD.4232;

      iD.4232 = 0;
      #pragma omp task shared(iD.4232)
        {
          {
            {
              iD.4232 = 1;
            }
          }
        }
      __builtin_GOMP_taskwaitD.494 ();
      #pragma omp parallel
        {
          {
            #pragma omp for nowait
            for (iD.4232 = 1; iD.4232 <= 8; iD.4232 = iD.4232 + 1)
              {
                {
                  fD.4229 (&iD.4232);
                }
                L.1D.4233:;
              }
          }
        }
    }
    [...]

Given this "early 'TREE_ADDRESSABLE'" (notice 'fD.4229 (&iD.4232);'), already
the gimplifier replaces the original loop variable 'iD.4232' with a
non-addressable 'i.1D.4239'.  Relevant part (edited) of '*.original' vs.
'*.gimple' dumps diff:

           #pragma omp parallel
             {
    +          integer(kind=4)D.8 i.1D.4239;
    +
    -        #pragma omp for nowait
    -        for (iD.4232 = 1; iD.4232 <= 8; iD.4232 = iD.4232 + 1)
    +          #pragma omp for nowait private(i.1D.4239) private(iD.4232)
    +          for (i.1D.4239 = 1; i.1D.4239 <= 8; i.1D.4239 = i.1D.4239 + 1)
                 {
    +              iD.4232 = i.1D.4239;
                   {
                     fD.4229 (&iD.4232);
                   }

In particular, notice that now a 'private(i.1D.4239)' appears in addition to
'private(iD.4232)'.

The OMP lowering then again replaces the loop variable 'i.1D.4239' with
'i.1D.4255' (I have not researched why) -- but it doesn't update the
'private(i.1D.4239)' clause accordingly (so that one persists, continues to
point to the original, now-unused DECL), and/or it doesn't introduce a new
'private(i.1D.4255)' replacement clause.  Similar for the 'iD.4232' ->
'iD.4256' transformation and unchanged 'private(iD.4232)' clause.  Relevant
part (edited) of '*.gimple' vs. '*.omplower' dumps diff:

           #pragma omp parallel [...]
           {
    -          integer(kind=4)D.8 i.1D.4239;
    +              integer(kind=4)D.8 iD.4256;
    +              integer(kind=4)D.8 i.1D.4255;

                   #pragma omp for nowait private(i.1D.4239) private(iD.4232)
    -          for (i.1D.4239 = 1; i.1D.4239 <= 8; i.1D.4239 = i.1D.4239 + 1)
    +              for (i.1D.4255 = 1; i.1D.4255 <= 8; i.1D.4255 = i.1D.4255 +
1)
                   {
    -              iD.4232 = i.1D.4239;
    +              iD.4256 = i.1D.4255;
                   {
    -                fD.4229 (&iD.4232);
    +                fD.4229 (&iD.4256); [static-chain: [...]]

I have confirmed that 'lookup_decl([i.1D.4239], ctx)' -> 'i.1D.4255' and
'lookup_decl([iD.4232], ctx)' -> 'iD.4256'.  The underlying assumption must be
that any such later use of the 'OMP_CLAUSE_DECL' of a 'OMP_CLAUSE_PRIVATE' is
filtered through 'lookup_decl'.  In other words, I assume it is intentional
that the 'OMP_CLAUSE_DECL' of a 'OMP_CLAUSE_PRIVATE' continues to point to the
original DECL, presumably because we still need to look up information from
that one?

Now, changing the scenario to the 'integer, value :: v' variant, the 'call
f(i)' is by-value, and 'i' is not made "early 'TREE_ADDRESSABLE'".

Relevant part (edited) of by-reference vs. by-value '*.original' dumps diff:

    -              fD.4229 (&iD.4232);
    +              fD.4229 (iD.4232);

Relevant part (edited) of by-reference vs. by-value '*.gimple' dumps diff:

    -          integer(kind=4)D.8 i.1D.4239;
    -
    -          #pragma omp for nowait private(i.1D.4239) private(iD.4232)
    -          for (i.1D.4239 = 1; i.1D.4239 <= 8; i.1D.4239 = i.1D.4239 + 1)
    +      #pragma omp for nowait private(iD.4232)
    +      for (iD.4232 = 1; iD.4232 <= 8; iD.4232 = iD.4232 + 1)
                 {
    -              iD.4232 = i.1D.4239;
                   {
    -                fD.4229 (&iD.4232);
    +            fD.4229 (iD.4232);

As 'iD.4232' itself is non-addressable at this time, the gimplifier doesn't
here introduce a non-addressable loop variable and corresponding 'private'
clause.

As you (Jakub) have explained above, it's now on the OMP lowering to make
'iD.4232' "late 'TREE_ADDRESSABLE'", but in the loop body, the loop variable
'iD.4254' can directly be used in the 'fD.4229 (iD.4254);' function call. 
Relevant part of '*.omplower' dump:

    [...]
        #pragma omp parallel shared(FRAME.1D.4239) [child fn:
MAIN__._omp_fn.1D.4249 (.omp_data_o.6D.4259)]
          {
            .omp_data_iD.4251 = (struct .omp_data_s.4D.4248 & restrict)
&.omp_data_o.6D.4259;
            {
              integer(kind=4)D.8 iD.4254;

              #pragma omp for nowait private(iD.4232)
              for (iD.4254 = 1; iD.4254 <= 8; iD.4254 = iD.4254 + 1)
              {
                D.4258 = .omp_data_iD.4251->FRAME.1D.4252;
                fD.4229 (iD.4254); [static-chain: D.4258]
    [...]

Again I have confirmed that 'lookup_decl([iD.4232], ctx)' -> 'iD.4254'.

And now, I understand the problem regarding the OpenACC privatization is that
it does the analysis ('gcc/omp-low.cc:oacc_privatization_scan_clause_chain'
etc.) on the original DECL, but does the transformation
('gcc/omp-low.cc:lower_oacc_private_marker' etc.) on the 'lookup_decl'ed DECL
-- and these need not correspond regarding 'TREE_ADDRESSABLE', as we've seen
above.

Now I'm looking into at which end to fix this: analysis vs. transformation,
and/or whether loop variables should altogether be excempt from this
transformation (but that may actually fall out of the former).

[Bug fortran/102330] [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932 since r12-980-g29a2f51806c

[cvs-commit at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

--- Comment #12 from CVS Commits <cvs-commit at gcc dot gnu.org> ---
The master branch has been updated by Thomas Schwinge <tschwinge@gcc.gnu.org>:

https://gcc.gnu.org/g:687091257820f4a6a005186437917270ecd27416

commit r12-7578-g687091257820f4a6a005186437917270ecd27416
Author: Thomas Schwinge <thomas@codesourcery.com>
Date:   Thu Jan 27 14:17:28 2022 +0100

    Add 'gfortran.dg/goacc-gomp/pr102330-{1,2,3}.f90' [PR102330]

    ..., currently XFAILed with 'dg-ice'.

            PR middle-end/102330
            gcc/testsuite/
            * gfortran.dg/goacc-gomp/pr102330-1.f90: New file.
            * gfortran.dg/goacc-gomp/pr102330-2.f90: Likewise.
            * gfortran.dg/goacc-gomp/pr102330-3.f90: Likewise.

[Bug fortran/102330] [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932 since r12-980-g29a2f51806c

[cvs-commit at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

--- Comment #13 from CVS Commits <cvs-commit at gcc dot gnu.org> ---
The master branch has been updated by Thomas Schwinge <tschwinge@gcc.gnu.org>:

https://gcc.gnu.org/g:7a5e036b61aa088e6b8564bc9383d37dfbb4801e

commit r12-7580-g7a5e036b61aa088e6b8564bc9383d37dfbb4801e
Author: Thomas Schwinge <thomas@codesourcery.com>
Date:   Mon Feb 14 16:56:35 2022 +0100

    [OpenACC privatization] Analyze 'lookup_decl'-translated DECL [PR90115,
PR102330, PR104774]

    ... so that it matches what we analyze and what we action on.
    Fix-up for commit 29a2f51806c5b30e17a8d0e9ba7915a3c53c34ff "openacc:
    Add support for gang local storage allocation in shared memory [PR90115]".

            PR middle-end/90115
            PR middle-end/102330
            PR middle-end/104774
            gcc/
            * omp-low.cc (oacc_privatization_candidate_p)
            (oacc_privatization_scan_clause_chain)
            (oacc_privatization_scan_decl_chain, lower_oacc_private_marker):
            Analyze 'lookup_decl'-translated DECL.
            gcc/testsuite/
            * c-c++-common/goacc/kernels-decompose-pr104061-1-3.c: Adjust.
            * c-c++-common/goacc/kernels-decompose-pr104061-1-4.c: Likewise.
            * c-c++-common/goacc/kernels-decompose-pr104132-1.c: Likewise.
            * c-c++-common/goacc/kernels-decompose-pr104133-1.c: Likewise.
            * c-c++-common/goacc/kernels-decompose-pr104774-1.c: Likewise.
            * c-c++-common/goacc/privatization-1-compute-loop.c: Likewise.
            * c-c++-common/goacc/privatization-1-compute.c: Likewise.
            * c-c++-common/goacc/privatization-1-routine_gang-loop.c:
            Likewise.
            * c-c++-common/goacc/privatization-1-routine_gang.c: Likewise.
            * gfortran.dg/goacc-gomp/pr102330-1.f90: Likewise, and subsume...
            * gfortran.dg/goacc-gomp/pr102330-2.f90: ... this file, and...
            * gfortran.dg/goacc-gomp/pr102330-3.f90: ... this file.
            * gfortran.dg/goacc/privatization-1-compute-loop.f90: Adjust.
            * gfortran.dg/goacc/privatization-1-compute.f90: Likewise.
            * gfortran.dg/goacc/privatization-1-routine_gang-loop.f90:
            Likewise.
            * gfortran.dg/goacc/privatization-1-routine_gang.f90: Likewise.
            libgomp/
            * testsuite/libgomp.oacc-c-c++-common/kernels-decompose-1.c:
            Enhance.
            *
testsuite/libgomp.oacc-c-c++-common/kernels-private-vars-local-worker-1.c:
            Adjust.
            *
testsuite/libgomp.oacc-c-c++-common/kernels-private-vars-local-worker-2.c:
            Likewise.
            *
testsuite/libgomp.oacc-c-c++-common/kernels-private-vars-local-worker-3.c:
            Likewise.
            *
testsuite/libgomp.oacc-c-c++-common/kernels-private-vars-local-worker-4.c:
            Likewise.
            *
testsuite/libgomp.oacc-c-c++-common/kernels-private-vars-local-worker-5.c:
            Likewise.
            *
testsuite/libgomp.oacc-c-c++-common/kernels-private-vars-loop-vector-1.c:
            Likewise.
            *
testsuite/libgomp.oacc-c-c++-common/kernels-private-vars-loop-vector-2.c:
            Likewise.
            *
testsuite/libgomp.oacc-c-c++-common/kernels-private-vars-loop-worker-2.c:
            Likewise.
            *
testsuite/libgomp.oacc-c-c++-common/kernels-private-vars-loop-worker-3.c:
            Likewise.
            *
testsuite/libgomp.oacc-c-c++-common/kernels-private-vars-loop-worker-4.c:
            Likewise.
            *
testsuite/libgomp.oacc-c-c++-common/kernels-private-vars-loop-worker-5.c:
            Likewise.
            *
testsuite/libgomp.oacc-c-c++-common/kernels-private-vars-loop-worker-6.c:
            Likewise.
            *
testsuite/libgomp.oacc-c-c++-common/kernels-private-vars-loop-worker-7.c:
            Likewise.
            * testsuite/libgomp.oacc-fortran/optional-private.f90: Likewise.
            * testsuite/libgomp.oacc-fortran/privatized-ref-1.f95: Likewise.
            * testsuite/libgomp.oacc-fortran/privatized-ref-2.f90: Likewise.

[Bug middle-end/102330] [12 Regression] ICE in expand_gimple_stmt_1, at cfgexpand.c:3932 since r12-980-g29a2f51806c

[tschwinge at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=102330

Thomas Schwinge <tschwinge at gcc dot gnu.org> changed:

           What    |Removed                     |Added
----------------------------------------------------------------------------
           See Also|                            |https://gcc.gnu.org/bugzill
                   |                            |a/show_bug.cgi?id=90115
          Component|fortran                     |middle-end
         Resolution|---                         |FIXED
             Status|ASSIGNED                    |RESOLVED

--- Comment #14 from Thomas Schwinge <tschwinge at gcc dot gnu.org> ---
.