| Message ID | alpine.LSU.2.20.1701250950500.12993@r111.fhfr.qr |
|---|---|
| State | New |
| Headers | show |
On Wed, Jan 25, 2017 at 09:52:41AM +0100, Richard Biener wrote: > 2017-01-25 Richard Biener <rguenther@suse.de> > > PR debug/78363 > * omp-expand.c: Include debug.h. > (expand_omp_taskreg): Make sure to generate early debug before > outlining anything from a function. > (expand_omp_target): Likewise. > (grid_expand_target_grid_body): Likewise. > > * g++.dg/gomp/pr78363-1.C: New testcase. > * g++.dg/gomp/pr78363-2.C: Likewise. > * g++.dg/gomp/pr78363-3.C: Likewise. Ok, with minor nit: > --- gcc/testsuite/g++.dg/gomp/pr78363-1.C (nonexistent) > +++ gcc/testsuite/g++.dg/gomp/pr78363-1.C (working copy) > @@ -0,0 +1,14 @@ > +// { dg-do compile } > +// { dg-require-effective-target c++11 } > +// { dg-options "-g -fopenmp" } > + > +int main() > +{ > + int n = 0; > + > +#pragma omp parallel for reduction (+: n) > + for (int i = [](){ return 3; }(); i < 10; ++i) > + n++; > + > + return n; > +} > Index: gcc/testsuite/g++.dg/gomp/pr78363-2.C > =================================================================== > --- gcc/testsuite/g++.dg/gomp/pr78363-2.C (nonexistent) > +++ gcc/testsuite/g++.dg/gomp/pr78363-2.C (working copy) > @@ -0,0 +1,15 @@ > +// { dg-do compile } > +// { dg-require-effective-target c++11 } > +// { dg-options "-g -fopenmp" } Please replace dg-options with: // { dg-additional-options "-g" } -fopenmp -Wno-hsa is the default, while dg-options of -g -fopenmp overrides that and -Wno-hsa would be lost. While it doesn't matter in the first and last testcase (no offloading in those), on this one I bet -Whsa (on by default) will warn if gcc is configured with hsa offloading, because it is not gridifiable. > + > +int main() > +{ > + int n = 0; > +#pragma omp target map(tofrom:n) > +#pragma omp for reduction (+: n) > + for (int i = [](){ return 3; }(); i < 10; ++i) > + n++; > + if (n != 7) > + __builtin_abort (); > + return 0; > +} > Index: gcc/testsuite/g++.dg/gomp/pr78363-3.C > =================================================================== > --- gcc/testsuite/g++.dg/gomp/pr78363-3.C (nonexistent) > +++ gcc/testsuite/g++.dg/gomp/pr78363-3.C (working copy) > @@ -0,0 +1,14 @@ > +// { dg-do compile } > +// { dg-require-effective-target c++11 } > +// { dg-options "-g -fopenmp" } > + > +int main() > +{ > + int n = 0; > +#pragma omp task shared (n) > + for (int i = [](){ return 3; }(); i < 10; ++i) > + n = i; > +#pragma omp taskwait > + if (n != 7) > + __builtin_abort (); > +} Jakub
Index: gcc/omp-expand.c =================================================================== --- gcc/omp-expand.c (revision 244890) +++ gcc/omp-expand.c (working copy) @@ -57,6 +57,7 @@ along with GCC; see the file COPYING3. #include "gomp-constants.h" #include "gimple-pretty-print.h" #include "hsa-common.h" +#include "debug.h" /* OMP region information. Every parallel and workshare @@ -1305,6 +1306,11 @@ expand_omp_taskreg (struct omp_region *r else block = gimple_block (entry_stmt); + /* Make sure to generate early debug for the function before + outlining anything. */ + if (! gimple_in_ssa_p (cfun)) + (*debug_hooks->early_global_decl) (cfun->decl); + new_bb = move_sese_region_to_fn (child_cfun, entry_bb, exit_bb, block); if (exit_bb) single_succ_edge (new_bb)->flags = EDGE_FALLTHRU; @@ -7016,6 +7022,11 @@ expand_omp_target (struct omp_region *re gsi_remove (&gsi, true); } + /* Make sure to generate early debug for the function before + outlining anything. */ + if (! gimple_in_ssa_p (cfun)) + (*debug_hooks->early_global_decl) (cfun->decl); + /* Move the offloading region into CHILD_CFUN. */ block = gimple_block (entry_stmt); @@ -7589,6 +7600,11 @@ grid_expand_target_grid_body (struct omp init_tree_ssa (cfun); pop_cfun (); + /* Make sure to generate early debug for the function before + outlining anything. */ + if (! gimple_in_ssa_p (cfun)) + (*debug_hooks->early_global_decl) (cfun->decl); + tree old_parm_decl = DECL_ARGUMENTS (kern_fndecl); gcc_assert (!DECL_CHAIN (old_parm_decl)); tree new_parm_decl = copy_node (DECL_ARGUMENTS (kern_fndecl)); Index: gcc/testsuite/g++.dg/gomp/pr78363-1.C =================================================================== --- gcc/testsuite/g++.dg/gomp/pr78363-1.C (nonexistent) +++ gcc/testsuite/g++.dg/gomp/pr78363-1.C (working copy) @@ -0,0 +1,14 @@ +// { dg-do compile } +// { dg-require-effective-target c++11 } +// { dg-options "-g -fopenmp" } + +int main() +{ + int n = 0; + +#pragma omp parallel for reduction (+: n) + for (int i = [](){ return 3; }(); i < 10; ++i) + n++; + + return n; +} Index: gcc/testsuite/g++.dg/gomp/pr78363-2.C =================================================================== --- gcc/testsuite/g++.dg/gomp/pr78363-2.C (nonexistent) +++ gcc/testsuite/g++.dg/gomp/pr78363-2.C (working copy) @@ -0,0 +1,15 @@ +// { dg-do compile } +// { dg-require-effective-target c++11 } +// { dg-options "-g -fopenmp" } + +int main() +{ + int n = 0; +#pragma omp target map(tofrom:n) +#pragma omp for reduction (+: n) + for (int i = [](){ return 3; }(); i < 10; ++i) + n++; + if (n != 7) + __builtin_abort (); + return 0; +} Index: gcc/testsuite/g++.dg/gomp/pr78363-3.C =================================================================== --- gcc/testsuite/g++.dg/gomp/pr78363-3.C (nonexistent) +++ gcc/testsuite/g++.dg/gomp/pr78363-3.C (working copy) @@ -0,0 +1,14 @@ +// { dg-do compile } +// { dg-require-effective-target c++11 } +// { dg-options "-g -fopenmp" } + +int main() +{ + int n = 0; +#pragma omp task shared (n) + for (int i = [](){ return 3; }(); i < 10; ++i) + n = i; +#pragma omp taskwait + if (n != 7) + __builtin_abort (); +}