Hi! On 2026-07-09T10:22:31+0200, Paul-Antoine Arras <[email protected]> wrote: > [...] attached patch.
Thanks, pushed in commit ab0dd7d5e16fcd90a58be84ec45a8a46994cf176 "openmp: Add OMPT entry points for worksharing-loop, single, and masked constructs". Grüße Thomas > From 7599a5edc9c806fcb14ae32d5b3b9bda0969ad50 Mon Sep 17 00:00:00 2001 > From: Paul-Antoine Arras <[email protected]> > Date: Tue, 30 Jun 2026 11:06:44 +0200 > Subject: [PATCH v4 1/2] openmp: Add OMPT entry points for worksharing-loop, > single, and masked constructs > > Add two new compiler flags, -fopenmp-ompt and -fopenmp-ompt-detailed, that > cause the compiler to emit OMPT-aware runtime entry points around OpenMP > worksharing constructs. With -fopenmp-ompt, the _start and _end variants > are called instead of the plain worksharing functions; with > -fopenmp-ompt-detailed, a _dispatch stub is additionally called after the > chunk bounds are computed. New stubs are added to libgomp as placeholders > for future OMPT callback invocations. > > gcc/ChangeLog: > > * common.opt (fopenmp-ompt): New option. > (fopenmp-ompt-detailed): New option. > * common.opt.urls: Regenerate. > * doc/invoke.texi (Option Summary): Add -fopenmp-ompt and > -fopenmp-ompt-detailed. > (OpenMP and OpenACC Options): Document -fopenmp-ompt and > -fopenmp-ompt-detailed. > * omp-builtins.def (BUILT_IN_GOMP_SINGLE_START_WITH_END): New builtin. > (BUILT_IN_GOMP_SINGLE_END): New builtin. > (BUILT_IN_GOMP_MASKED_END): New builtin. > (BUILT_IN_GOMP_LOOP_STATIC_WORKSHARING_START): New builtin. > (BUILT_IN_GOMP_LOOP_STATIC_WORKSHARING_END): New builtin. > (BUILT_IN_GOMP_LOOP_STATIC_WORKSHARING_DISPATCH): New builtin. > (BUILT_IN_GOMP_DISTRIBUTE_STATIC_WORKSHARING_START): New builtin. > (BUILT_IN_GOMP_DISTRIBUTE_STATIC_WORKSHARING_END): New builtin. > (BUILT_IN_GOMP_DISTRIBUTE_STATIC_WORKSHARING_DISPATCH): New builtin. > * omp-expand.cc (expand_omp_for_static_nochunk): With -fopenmp-ompt, > emit _start and _end variants instead of the plain worksharing call; > with -fopenmp-ompt-detailed, also emit _dispatch. > (expand_omp_for_static_chunk): Likewise. > * omp-low.cc (lower_omp_single_simple): With -fopenmp-ompt, use > GOMP_single_start_with_end instead of GOMP_single_start. > (lower_omp_single): With -fopenmp-ompt, emit GOMP_single_end. > (lower_omp_master): With -fopenmp-ompt, emit GOMP_masked_end. > * opts.cc (finish_options): Diagnose -fopenmp-ompt and > -fopenmp-ompt-detailed used without -fopenmp. > > libgomp/ChangeLog: > > * config/gcn/teams.c (GOMP_distribute_static_worksharing_start, > GOMP_distribute_static_worksharing_dispatch, > GOMP_distribute_static_worksharing_end): New functions. > * config/nvptx/teams.c (GOMP_distribute_static_worksharing_start, > GOMP_distribute_static_worksharing_dispatch, > GOMP_distribute_static_worksharing_end): Likewise. > * libgomp.map (GOMP_6.0.2): Export new entry points. > * libgomp_g.h: Add prototypes for GOMP_loop_static_worksharing, > GOMP_loop_static_worksharing_start, > GOMP_loop_static_worksharing_dispatch, > GOMP_loop_static_worksharing_end, GOMP_has_masked_thread_num, > GOMP_masked_end, GOMP_single_start_with_end, GOMP_single_end, > GOMP_distribute_static_worksharing_start, > GOMP_distribute_static_worksharing_dispatch, and > GOMP_distribute_static_worksharing_end. > * loop.c (GOMP_loop_static_worksharing): Move here from parallel.c. > (GOMP_loop_static_worksharing_start): New function. > (GOMP_loop_static_worksharing_dispatch): New stub. > (GOMP_loop_static_worksharing_end): New stub. > (GOMP_has_masked_thread_num): Move here from parallel.c. > (GOMP_masked_end): New stub. > * parallel.c (GOMP_loop_static_worksharing): Move to loop.c. > (GOMP_has_masked_thread_num): Move to loop.c. > * single.c (GOMP_single_start_with_end): New function. > (GOMP_single_end): New stub. > (GOMP_single_copy_start): Fix trailing whitespace. > * teams.c (GOMP_distribute_static_worksharing_start): New function. > (GOMP_distribute_static_worksharing_dispatch): New stub. > (GOMP_distribute_static_worksharing_end): New stub. > * testsuite/libgomp.c-c++-common/for-static-1.c: New test. > * testsuite/libgomp.c-c++-common/for-static-2.c: New test. > * testsuite/libgomp.c-c++-common/for-static-3.c: New test. > * testsuite/libgomp.c-c++-common/for-static.h: New test. > > gcc/testsuite/ChangeLog: > > * c-c++-common/gomp/masked-1.c: Add scan-tree-dump-not check that > GOMP_masked_end is not emitted without -fopenmp-ompt. > * c-c++-common/gomp/for-8.c: Remove; superseded by for-static-*.c. > * c-c++-common/gomp/for-static-1.c: New test; checks that the plain > worksharing builtins are emitted without -fopenmp-ompt. > * c-c++-common/gomp/for-static-2.c: New test; checks that _start and > _end variants are emitted with -fopenmp-ompt. > * c-c++-common/gomp/for-static-3.c: New test; checks that _dispatch is > also emitted with -fopenmp-ompt-detailed. > * c-c++-common/gomp/for-static.h: New shared test header. > * c-c++-common/gomp/masked-3.c: New test for GOMP_masked_end emission > with -fopenmp-ompt. > * c-c++-common/gomp/openmp-ompt-1.c: New test; checks that > -fopenmp-ompt without -fopenmp is diagnosed. > * c-c++-common/gomp/openmp-ompt-2.c: New test; checks that > -fopenmp-ompt-detailed without -fopenmp is diagnosed. > * c-c++-common/gomp/single2.c: New test for GOMP_single_start_with_end > and GOMP_single_end emission with -fopenmp-ompt. > --- > gcc/common.opt | 8 ++ > gcc/common.opt.urls | 6 + > gcc/doc/invoke.texi | 21 +++- > gcc/omp-builtins.def | 25 ++++ > gcc/omp-expand.cc | 111 +++++++++++++++++- > gcc/omp-low.cc | 21 +++- > gcc/opts.cc | 6 + > gcc/testsuite/c-c++-common/gomp/for-8.c | 67 ----------- > .../c-c++-common/gomp/for-static-1.c | 17 +++ > .../c-c++-common/gomp/for-static-2.c | 15 +++ > .../c-c++-common/gomp/for-static-3.c | 14 +++ > gcc/testsuite/c-c++-common/gomp/for-static.h | 51 ++++++++ > gcc/testsuite/c-c++-common/gomp/masked-1.c | 1 + > gcc/testsuite/c-c++-common/gomp/masked-3.c | 15 +++ > .../c-c++-common/gomp/openmp-ompt-1.c | 4 + > .../c-c++-common/gomp/openmp-ompt-2.c | 4 + > gcc/testsuite/c-c++-common/gomp/single2.c | 13 ++ > libgomp/config/gcn/teams.c | 23 ++++ > libgomp/config/nvptx/teams.c | 21 ++++ > libgomp/libgomp.map | 9 ++ > libgomp/libgomp_g.h | 12 +- > libgomp/loop.c | 35 ++++++ > libgomp/parallel.c | 18 +-- > libgomp/single.c | 33 +++++- > libgomp/teams.c | 23 ++++ > .../libgomp.c-c++-common/for-static-1.c | 13 ++ > .../libgomp.c-c++-common/for-static-2.c | 11 ++ > .../libgomp.c-c++-common/for-static-3.c | 11 ++ > .../libgomp.c-c++-common/for-static.h | 37 ++++++ > 29 files changed, 556 insertions(+), 89 deletions(-) > delete mode 100644 gcc/testsuite/c-c++-common/gomp/for-8.c > create mode 100644 gcc/testsuite/c-c++-common/gomp/for-static-1.c > create mode 100644 gcc/testsuite/c-c++-common/gomp/for-static-2.c > create mode 100644 gcc/testsuite/c-c++-common/gomp/for-static-3.c > create mode 100644 gcc/testsuite/c-c++-common/gomp/for-static.h > create mode 100644 gcc/testsuite/c-c++-common/gomp/masked-3.c > create mode 100644 gcc/testsuite/c-c++-common/gomp/openmp-ompt-1.c > create mode 100644 gcc/testsuite/c-c++-common/gomp/openmp-ompt-2.c > create mode 100644 gcc/testsuite/c-c++-common/gomp/single2.c > create mode 100644 libgomp/testsuite/libgomp.c-c++-common/for-static-1.c > create mode 100644 libgomp/testsuite/libgomp.c-c++-common/for-static-2.c > create mode 100644 libgomp/testsuite/libgomp.c-c++-common/for-static-3.c > create mode 100644 libgomp/testsuite/libgomp.c-c++-common/for-static.h > > diff --git a/gcc/common.opt b/gcc/common.opt > index 218dddf5dfe..148d0ac3d17 100644 > --- a/gcc/common.opt > +++ b/gcc/common.opt > @@ -2503,6 +2503,14 @@ fomit-frame-pointer > Common Var(flag_omit_frame_pointer) Optimization > When possible do not generate stack frames. > > +fopenmp-ompt > +Common Var(flag_openmp_ompt) EnabledBy(fopenmp-ompt-detailed) > +Emit additional calls into libgomp, enabling OMPT callbacks before and after > some OpenMP worksharing constructs. > + > +fopenmp-ompt-detailed > +Common Var(flag_openmp_ompt_detailed) > +Like -fopenmp-ompt, and emit additional calls into libgomp, enabling OMPT > dispatch callbacks. > + > fopenmp-target-simd-clone > Common Alias(fopenmp-target-simd-clone=,any,none) > > diff --git a/gcc/common.opt.urls b/gcc/common.opt.urls > index 490146dde8c..567a8dfc2d6 100644 > --- a/gcc/common.opt.urls > +++ b/gcc/common.opt.urls > @@ -1336,6 +1336,12 @@ > UrlSuffix(gcc/OpenMP-and-OpenACC-Options.html#index-foffload-options) > fomit-frame-pointer > UrlSuffix(gcc/Optimize-Options.html#index-fno-omit-frame-pointer) > > +fopenmp-ompt > +UrlSuffix(gcc/OpenMP-and-OpenACC-Options.html#index-fno-openmp-ompt) > + > +fopenmp-ompt-detailed > +UrlSuffix(gcc/OpenMP-and-OpenACC-Options.html#index-fno-openmp-ompt-detailed) > + > fopenmp-target-simd-clone > > UrlSuffix(gcc/OpenMP-and-OpenACC-Options.html#index-fno-openmp-target-simd-clone) > > diff --git a/gcc/doc/invoke.texi b/gcc/doc/invoke.texi > index 4fbd25e49dd..d1cbfdbc7a7 100644 > --- a/gcc/doc/invoke.texi > +++ b/gcc/doc/invoke.texi > @@ -323,7 +323,8 @@ Objective-C and Objective-C++ Dialects}. > @xref{OpenMP and OpenACC Options,,Options Controlling OpenMP and OpenACC}. > @gccoptlist{-foffload=@var{arg} -foffload-options=@var{arg} > -fopenacc -fopenacc-dim=@var{geom} > --fopenmp -fopenmp-simd > -fopenmp-target-simd-clone@r{[}=@var{device-type}@r{]}} > +-fopenmp -fopenmp-simd -fopenmp-ompt -fopenmp-ompt-detailed > +-fopenmp-target-simd-clone@r{[}=@var{device-type}@r{]}} > > @item Diagnostic Message Formatting Options > @xref{Diagnostic Message Formatting Options,,Options to Control Diagnostic > Messages Formatting}. > @@ -5600,6 +5601,24 @@ Fortran. In fixed source form Fortran, the sentinels > can also start with > @option{-fopenmp} is additionally specified, the @code{loop} region binds > to the current task region, independent of the specified @code{bind} clause. > > +@opindex fopenmp-ompt > +@opindex fno-openmp-ompt > +@cindex OpenMP OMPT > +@item -fopenmp-ompt > +Emit additional calls into libgomp, enabling OMPT callbacks around some > OpenMP > +worksharing constructs (namely statically scheduled @code{for} and > +@code{distribute}, @code{single}, @code{masked} and @code{master}), so that > an > +OMPT tool can observe their start and end. Requires @option{-fopenmp}. > + > +@opindex fopenmp-ompt-detailed > +@opindex fno-openmp-ompt-detailed > +@cindex OpenMP OMPT > +@item -fopenmp-ompt-detailed > +Like @option{-fopenmp-ompt}, and additionally emit calls into libgomp, > enabling > +OMPT dispatch callbacks that get called at the beginning of each workshare > chunk > +for statically scheduled @code{for} and @code{distribute}. Requires > +@option{-fopenmp}. > + > @opindex fopenmp-target-simd-clone > @opindex fno-openmp-target-simd-clone > @cindex OpenMP target SIMD clone > diff --git a/gcc/omp-builtins.def b/gcc/omp-builtins.def > index d7440a42761..465aeebefdb 100644 > --- a/gcc/omp-builtins.def > +++ b/gcc/omp-builtins.def > @@ -445,6 +445,11 @@ DEF_GOMP_BUILTIN (BUILT_IN_GOMP_SECTIONS_END_NOWAIT, > BT_FN_VOID, ATTR_NOTHROW_LEAF_LIST) > DEF_GOMP_BUILTIN (BUILT_IN_GOMP_SINGLE_START, "GOMP_single_start", > BT_FN_BOOL, ATTR_NOTHROW_LEAF_LIST) > +DEF_GOMP_BUILTIN (BUILT_IN_GOMP_SINGLE_START_WITH_END, > + "GOMP_single_start_with_end", BT_FN_BOOL, > + ATTR_NOTHROW_LEAF_LIST) > +DEF_GOMP_BUILTIN (BUILT_IN_GOMP_SINGLE_END, "GOMP_single_end", BT_FN_VOID, > + ATTR_NOTHROW_LEAF_LIST) > DEF_GOMP_BUILTIN (BUILT_IN_GOMP_SINGLE_COPY_START, "GOMP_single_copy_start", > BT_FN_PTR, ATTR_NOTHROW_LEAF_LIST) > DEF_GOMP_BUILTIN (BUILT_IN_GOMP_SINGLE_COPY_END, "GOMP_single_copy_end", > @@ -503,9 +508,29 @@ DEF_GOMP_BUILTIN (BUILT_IN_GOMP_ERROR, "GOMP_error", > DEF_GOMP_BUILTIN (BUILT_IN_GOMP_HAS_MASKED_THREAD_NUM, > "GOMP_has_masked_thread_num", BT_FN_BOOL_INT, > ATTR_CONST_NOTHROW_LEAF_LIST) > +DEF_GOMP_BUILTIN (BUILT_IN_GOMP_MASKED_END, "GOMP_masked_end", BT_FN_VOID, > + ATTR_NOTHROW_LEAF_LIST) > DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_STATIC_WORKSHARING, > "GOMP_loop_static_worksharing", BT_FN_COMPLEX_INT, > ATTR_CONST_NOTHROW_LEAF_LIST) > +DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_STATIC_WORKSHARING_START, > + "GOMP_loop_static_worksharing_start", BT_FN_COMPLEX_INT, > + ATTR_CONST_NOTHROW_LEAF_LIST) > +DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_STATIC_WORKSHARING_END, > + "GOMP_loop_static_worksharing_end", BT_FN_VOID, > + ATTR_NOTHROW_LEAF_LIST) > +DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_STATIC_WORKSHARING_DISPATCH, > + "GOMP_loop_static_worksharing_dispatch", BT_FN_VOID, > + ATTR_NOTHROW_LEAF_LIST) > DEF_GOMP_BUILTIN (BUILT_IN_GOMP_DISTRIBUTE_STATIC_WORKSHARING, > "GOMP_distribute_static_worksharing", BT_FN_COMPLEX_INT, > ATTR_CONST_NOTHROW_LEAF_LIST) > +DEF_GOMP_BUILTIN (BUILT_IN_GOMP_DISTRIBUTE_STATIC_WORKSHARING_START, > + "GOMP_distribute_static_worksharing_start", BT_FN_COMPLEX_INT, > + ATTR_CONST_NOTHROW_LEAF_LIST) > +DEF_GOMP_BUILTIN (BUILT_IN_GOMP_DISTRIBUTE_STATIC_WORKSHARING_END, > + "GOMP_distribute_static_worksharing_end", BT_FN_VOID, > + ATTR_NOTHROW_LEAF_LIST) > +DEF_GOMP_BUILTIN (BUILT_IN_GOMP_DISTRIBUTE_STATIC_WORKSHARING_DISPATCH, > + "GOMP_distribute_static_worksharing_dispatch", BT_FN_VOID, > + ATTR_NOTHROW_LEAF_LIST) > diff --git a/gcc/omp-expand.cc b/gcc/omp-expand.cc > index 3d799d28032..ad380b27d06 100644 > --- a/gcc/omp-expand.cc > +++ b/gcc/omp-expand.cc > @@ -5201,10 +5201,14 @@ expand_omp_for_static_nochunk (struct omp_region > *region, > switch (gimple_omp_for_kind (fd->for_stmt)) > { > case GF_OMP_FOR_KIND_FOR: > - decl = builtin_decl_explicit (BUILT_IN_GOMP_LOOP_STATIC_WORKSHARING); > + decl = builtin_decl_explicit ( > + flag_openmp_ompt ? BUILT_IN_GOMP_LOOP_STATIC_WORKSHARING_START > + : BUILT_IN_GOMP_LOOP_STATIC_WORKSHARING); > break; > case GF_OMP_FOR_KIND_DISTRIBUTE: > - decl = builtin_decl_explicit > (BUILT_IN_GOMP_DISTRIBUTE_STATIC_WORKSHARING); > + decl = builtin_decl_explicit ( > + flag_openmp_ompt ? BUILT_IN_GOMP_DISTRIBUTE_STATIC_WORKSHARING_START > + : BUILT_IN_GOMP_DISTRIBUTE_STATIC_WORKSHARING); > break; > default: > gcc_unreachable (); > @@ -5576,8 +5580,53 @@ expand_omp_for_static_nochunk (struct omp_region > *region, > cont_bb, body_bb); > } > > - /* Replace the GIMPLE_OMP_RETURN with a barrier, or nothing. */ > + if (flag_openmp_ompt_detailed) > + { > + /* Insert call to GOMP_*_static_worksharing_dispatch at the end of > + seq_start_bb. */ > + gsi = gsi_last_nondebug_bb (seq_start_bb); > + tree decl; > + switch (gimple_omp_for_kind (fd->for_stmt)) > + { > + case GF_OMP_FOR_KIND_FOR: > + decl = builtin_decl_explicit ( > + BUILT_IN_GOMP_LOOP_STATIC_WORKSHARING_DISPATCH); > + break; > + case GF_OMP_FOR_KIND_DISTRIBUTE: > + decl = builtin_decl_explicit ( > + BUILT_IN_GOMP_DISTRIBUTE_STATIC_WORKSHARING_DISPATCH); > + break; > + default: > + gcc_unreachable (); > + } > + gcall *g = gimple_build_call (decl, 0); > + gsi_insert_before (&gsi, g, GSI_SAME_STMT); > + } > + > gsi = gsi_last_nondebug_bb (exit_bb); > + if (flag_openmp_ompt) > + { > + /* Insert call to GOMP_*_static_worksharing_end at the end of exit_bb. > + */ > + tree decl; > + switch (gimple_omp_for_kind (fd->for_stmt)) > + { > + case GF_OMP_FOR_KIND_FOR: > + decl > + = builtin_decl_explicit (BUILT_IN_GOMP_LOOP_STATIC_WORKSHARING_END); > + break; > + case GF_OMP_FOR_KIND_DISTRIBUTE: > + decl = builtin_decl_explicit ( > + BUILT_IN_GOMP_DISTRIBUTE_STATIC_WORKSHARING_END); > + break; > + default: > + gcc_unreachable (); > + } > + gcall *g = gimple_build_call (decl, 0); > + gsi_insert_after (&gsi, g, GSI_SAME_STMT); > + } > + > + /* Replace the GIMPLE_OMP_RETURN with a barrier, or nothing. */ > if (!gimple_omp_return_nowait_p (gsi_stmt (gsi))) > { > t = gimple_omp_return_lhs (gsi_stmt (gsi)); > @@ -5646,6 +5695,7 @@ expand_omp_for_static_nochunk (struct omp_region > *region, > exit3_bb = split_block (exit2_bb, g)->dest; > gsi = gsi_after_labels (exit3_bb); > } > + > gsi_remove (&gsi, true); > > /* Connect all the blocks. */ > @@ -5969,10 +6019,14 @@ expand_omp_for_static_chunk (struct omp_region > *region, > switch (gimple_omp_for_kind (fd->for_stmt)) > { > case GF_OMP_FOR_KIND_FOR: > - decl = builtin_decl_explicit (BUILT_IN_GOMP_LOOP_STATIC_WORKSHARING); > + decl = builtin_decl_explicit ( > + flag_openmp_ompt ? BUILT_IN_GOMP_LOOP_STATIC_WORKSHARING_START > + : BUILT_IN_GOMP_LOOP_STATIC_WORKSHARING); > break; > case GF_OMP_FOR_KIND_DISTRIBUTE: > - decl = builtin_decl_explicit > (BUILT_IN_GOMP_DISTRIBUTE_STATIC_WORKSHARING); > + decl = builtin_decl_explicit ( > + flag_openmp_ompt ? BUILT_IN_GOMP_DISTRIBUTE_STATIC_WORKSHARING_START > + : BUILT_IN_GOMP_DISTRIBUTE_STATIC_WORKSHARING); > break; > default: > gcc_unreachable (); > @@ -6303,8 +6357,30 @@ expand_omp_for_static_chunk (struct omp_region *region, > gsi_insert_after (&gsi, assign_stmt, GSI_CONTINUE_LINKING); > } > > - /* Replace the GIMPLE_OMP_RETURN with a barrier, or nothing. */ > gsi = gsi_last_nondebug_bb (exit_bb); > + if (flag_openmp_ompt) > + { > + /* Insert call to GOMP_*_static_worksharing_end at the end of exit_bb. > + */ > + tree decl; > + switch (gimple_omp_for_kind (fd->for_stmt)) > + { > + case GF_OMP_FOR_KIND_FOR: > + decl > + = builtin_decl_explicit (BUILT_IN_GOMP_LOOP_STATIC_WORKSHARING_END); > + break; > + case GF_OMP_FOR_KIND_DISTRIBUTE: > + decl = builtin_decl_explicit ( > + BUILT_IN_GOMP_DISTRIBUTE_STATIC_WORKSHARING_END); > + break; > + default: > + gcc_unreachable (); > + } > + gcall *g = gimple_build_call (decl, 0); > + gsi_insert_after (&gsi, g, GSI_SAME_STMT); > + } > + > + /* Replace the GIMPLE_OMP_RETURN with a barrier, or nothing. */ > if (!gimple_omp_return_nowait_p (gsi_stmt (gsi))) > { > t = gimple_omp_return_lhs (gsi_stmt (gsi)); > @@ -6340,6 +6416,29 @@ expand_omp_for_static_chunk (struct omp_region *region, > } > gsi_remove (&gsi, true); > > + if (flag_openmp_ompt_detailed) > + { > + /* Insert call to GOMP_*_static_worksharing_dispatch at the end of > + seq_start_bb. */ > + gsi = gsi_last_nondebug_bb (seq_start_bb); > + tree decl; > + switch (gimple_omp_for_kind (fd->for_stmt)) > + { > + case GF_OMP_FOR_KIND_FOR: > + decl = builtin_decl_explicit ( > + BUILT_IN_GOMP_LOOP_STATIC_WORKSHARING_DISPATCH); > + break; > + case GF_OMP_FOR_KIND_DISTRIBUTE: > + decl = builtin_decl_explicit ( > + BUILT_IN_GOMP_DISTRIBUTE_STATIC_WORKSHARING_DISPATCH); > + break; > + default: > + gcc_unreachable (); > + } > + gcall *g = gimple_build_call (decl, 0); > + gsi_insert_before (&gsi, g, GSI_SAME_STMT); > + } > + > /* Connect the new blocks. */ > find_edge (iter_part_bb, seq_start_bb)->flags = EDGE_TRUE_VALUE; > find_edge (iter_part_bb, fin_bb)->flags = EDGE_FALSE_VALUE; > diff --git a/gcc/omp-low.cc b/gcc/omp-low.cc > index dbdbc4079f4..50da8581a8d 100644 > --- a/gcc/omp-low.cc > +++ b/gcc/omp-low.cc > @@ -8786,7 +8786,9 @@ lower_omp_single_simple (gomp_single *single_stmt, > gimple_seq *pre_p) > gimple *call, *cond; > tree lhs, decl; > > - decl = builtin_decl_explicit (BUILT_IN_GOMP_SINGLE_START); > + decl = builtin_decl_explicit (flag_openmp_ompt > + ? BUILT_IN_GOMP_SINGLE_START_WITH_END > + : BUILT_IN_GOMP_SINGLE_START); > lhs = create_tmp_var (TREE_TYPE (TREE_TYPE (decl))); > call = gimple_build_call (decl, 0); > gimple_call_set_lhs (call, lhs); > @@ -8923,6 +8925,15 @@ lower_omp_single (gimple_stmt_iterator *gsi_p, > omp_context *ctx) > gimple *g = gimple_build_omp_return (nowait); > gimple_seq_add_stmt (&bind_body_tail, g); > maybe_add_implicit_barrier_cancel (ctx, g, &bind_body_tail); > + > + if (flag_openmp_ompt && !ctx->record_type) > + { > + /* Insert call to GOMP_single_end. */ > + tree decl = builtin_decl_explicit (BUILT_IN_GOMP_SINGLE_END); > + gimple *stmt = gimple_build_call (decl, 0); > + gimple_seq_add_stmt (&bind_body_tail, stmt); > + } > + > if (ctx->record_type) > { > gimple_stmt_iterator gsi = gsi_start (bind_body_tail); > @@ -9099,6 +9110,14 @@ lower_omp_master (gimple_stmt_iterator *gsi_p, > omp_context *ctx) > gimple_bind_add_seq (bind, gimple_omp_body (stmt)); > gimple_omp_set_body (stmt, NULL); > > + if (flag_openmp_ompt) > + { > + /* Insert call to GOMP_masked_end at the end of the body. */ > + tree decl = builtin_decl_explicit (BUILT_IN_GOMP_MASKED_END); > + gcall *g = gimple_build_call (decl, 0); > + gimple_bind_add_stmt (bind, g); > + } > + > gimple_bind_add_stmt (bind, gimple_build_label (lab)); > > gimple_bind_add_stmt (bind, gimple_build_omp_return (true)); > diff --git a/gcc/opts.cc b/gcc/opts.cc > index 342517528e8..3dae57af0b4 100644 > --- a/gcc/opts.cc > +++ b/gcc/opts.cc > @@ -1538,6 +1538,12 @@ finish_options (struct gcc_options *opts, struct > gcc_options *opts_set, > " %<-fstrict-flex-arrays%> is not present"); > } > > + if ((opts->x_flag_openmp_ompt || opts->x_flag_openmp_ompt_detailed) > + && !opts->x_flag_openmp) > + error_at ( > + loc, > + "%<-fopenmp-ompt%> and %<-fopenmp-ompt-detailed%> require > %<-fopenmp%>"); > + > diagnose_options (opts, opts_set, loc); > } > > diff --git a/gcc/testsuite/c-c++-common/gomp/for-8.c > b/gcc/testsuite/c-c++-common/gomp/for-8.c > deleted file mode 100644 > index a894aaaaed1..00000000000 > --- a/gcc/testsuite/c-c++-common/gomp/for-8.c > +++ /dev/null > @@ -1,67 +0,0 @@ > -/* { dg-do compile } */ > -/* { dg-additional-options "-fdump-tree-ompexp" } */ > - > -/* Check that the static loop and distribute expanders fetch the thread/team > - id and count through a single GOMP_loop_static_worksharing / > - GOMP_distribute_static_worksharing call rather than separate > - omp_get_thread_num/omp_get_num_threads or > omp_get_team_num/omp_get_num_teams > - calls. */ > - > -void bar (int); > - > -/* Static schedule without a chunk size goes through > - expand_omp_for_static_nochunk. */ > - > -void > -f1 (int n) > -{ > - int i; > - #pragma omp for schedule(static) > - for (i = 0; i < n; ++i) > - bar (i); > -} > - > -/* Static schedule with a chunk size goes through > - expand_omp_for_static_chunk. */ > - > -void > -f2 (int n) > -{ > - int i; > - #pragma omp for schedule(static, 4) > - for (i = 0; i < n; ++i) > - bar (i); > -} > - > -/* Distribute without a chunk size goes through > - expand_omp_for_static_nochunk. */ > - > -void > -f3 (int n) > -{ > - int i; > -#pragma omp teams > -#pragma omp distribute > - for (i = 0; i < n; ++i) > - bar (i); > -} > - > -/* Distribute with a chunk size goes through > - expand_omp_for_static_chunk. */ > - > -void > -f4 (int n) > -{ > - int i; > -#pragma omp teams > -#pragma omp distribute dist_schedule(static, 4) > - for (i = 0; i < n; ++i) > - bar (i); > -} > - > -/* { dg-final { scan-tree-dump "GOMP_loop_static_worksharing" "ompexp" } } */ > -/* { dg-final { scan-tree-dump "GOMP_distribute_static_worksharing" "ompexp" > } } */ > -/* { dg-final { scan-tree-dump-not "omp_get_num_threads" "ompexp" } } */ > -/* { dg-final { scan-tree-dump-not "omp_get_thread_num" "ompexp" } } */ > -/* { dg-final { scan-tree-dump-not "omp_get_num_teams" "ompexp" } } */ > -/* { dg-final { scan-tree-dump-not "omp_get_team_num" "ompexp" } } */ > diff --git a/gcc/testsuite/c-c++-common/gomp/for-static-1.c > b/gcc/testsuite/c-c++-common/gomp/for-static-1.c > new file mode 100644 > index 00000000000..50d4e176f3d > --- /dev/null > +++ b/gcc/testsuite/c-c++-common/gomp/for-static-1.c > @@ -0,0 +1,17 @@ > +/* { dg-do compile } */ > +/* { dg-additional-options "-fdump-tree-ompexp" } */ > + > +/* Check that the static loop and distribute expanders fetch the thread/team > + id and count through a single GOMP_loop_static_worksharing / > + GOMP_distribute_static_worksharing call rather than separate > + omp_get_thread_num/omp_get_num_threads or > omp_get_team_num/omp_get_num_teams > + calls. */ > + > +#include "for-static.h" > + > +/* { dg-final { scan-tree-dump-times "GOMP_loop_static_worksharing \\(" 2 > "ompexp" } } */ > +/* { dg-final { scan-tree-dump-times "GOMP_distribute_static_worksharing > \\(" 2 "ompexp" } } */ > +/* { dg-final { scan-tree-dump-not "omp_get_num_threads" "ompexp" } } */ > +/* { dg-final { scan-tree-dump-not "omp_get_thread_num" "ompexp" } } */ > +/* { dg-final { scan-tree-dump-not "omp_get_num_teams" "ompexp" } } */ > +/* { dg-final { scan-tree-dump-not "omp_get_team_num" "ompexp" } } */ > diff --git a/gcc/testsuite/c-c++-common/gomp/for-static-2.c > b/gcc/testsuite/c-c++-common/gomp/for-static-2.c > new file mode 100644 > index 00000000000..c3d25b97bed > --- /dev/null > +++ b/gcc/testsuite/c-c++-common/gomp/for-static-2.c > @@ -0,0 +1,15 @@ > +/* { dg-do compile } */ > +/* { dg-additional-options "-fopenmp-ompt -fdump-tree-ompexp" } */ > + > +/* Check that, with -fopenmp-ompt, the _start and _end variants are called. > */ > + > +#include "for-static.h" > + > +/* { dg-final { scan-tree-dump-times "GOMP_loop_static_worksharing_start > \\(" 2 "ompexp" } } */ > +/* { dg-final { scan-tree-dump-times "GOMP_loop_static_worksharing_end \\(" > 2 "ompexp" } } */ > +/* { dg-final { scan-tree-dump-times > "GOMP_distribute_static_worksharing_start \\(" 2 "ompexp" } } */ > +/* { dg-final { scan-tree-dump-times "GOMP_distribute_static_worksharing_end > \\(" 2 "ompexp" } } */ > +/* { dg-final { scan-tree-dump-not "GOMP_loop_static_worksharing \\(" > "ompexp" } } */ > +/* { dg-final { scan-tree-dump-not "GOMP_loop_static_worksharing_dispatch" > "ompexp" } } */ > +/* { dg-final { scan-tree-dump-not "GOMP_distribute_static_worksharing \\(" > "ompexp" } } */ > +/* { dg-final { scan-tree-dump-not > "GOMP_distribute_static_worksharing_dispatch" "ompexp" } } */ > diff --git a/gcc/testsuite/c-c++-common/gomp/for-static-3.c > b/gcc/testsuite/c-c++-common/gomp/for-static-3.c > new file mode 100644 > index 00000000000..6fd1da6c407 > --- /dev/null > +++ b/gcc/testsuite/c-c++-common/gomp/for-static-3.c > @@ -0,0 +1,14 @@ > +/* { dg-do compile } */ > +/* { dg-additional-options "-fopenmp-ompt-detailed -fdump-tree-ompexp" } */ > + > +/* Check that, with -fopenmp-ompt-detailed, _dispatch builtins are called > along > + with the _start and _end variants. */ > + > +#include "for-static.h" > + > +/* { dg-final { scan-tree-dump-times "GOMP_loop_static_worksharing_start > \\(" 2 "ompexp" } } */ > +/* { dg-final { scan-tree-dump-times "GOMP_loop_static_worksharing_dispatch > \\(" 2 "ompexp" } } */ > +/* { dg-final { scan-tree-dump-times "GOMP_loop_static_worksharing_end \\(" > 2 "ompexp" } } */ > +/* { dg-final { scan-tree-dump-times > "GOMP_distribute_static_worksharing_start \\(" 2 "ompexp" } } */ > +/* { dg-final { scan-tree-dump-times > "GOMP_distribute_static_worksharing_dispatch \\(" 2 "ompexp" } } */ > +/* { dg-final { scan-tree-dump-times "GOMP_distribute_static_worksharing_end > \\(" 2 "ompexp" } } */ > diff --git a/gcc/testsuite/c-c++-common/gomp/for-static.h > b/gcc/testsuite/c-c++-common/gomp/for-static.h > new file mode 100644 > index 00000000000..6504c75a13e > --- /dev/null > +++ b/gcc/testsuite/c-c++-common/gomp/for-static.h > @@ -0,0 +1,51 @@ > +void bar (int); > + > +/* Static schedule without a chunk size goes through > + expand_omp_for_static_nochunk. */ > + > +void > +f1 (int n) > +{ > + int i; > + #pragma omp for schedule(static) > + for (i = 0; i < n; ++i) > + bar (i); > +} > + > +/* Static schedule with a chunk size goes through > + expand_omp_for_static_chunk. */ > + > +void > +f2 (int n) > +{ > + int i; > + #pragma omp for schedule(static, 4) > + for (i = 0; i < n; ++i) > + bar (i); > +} > + > +/* Distribute without a chunk size goes through > + expand_omp_for_static_nochunk. */ > + > +void > +f3 (int n) > +{ > + int i; > +#pragma omp teams > +#pragma omp distribute > + for (i = 0; i < n; ++i) > + bar (i); > +} > + > +/* Distribute with a chunk size goes through > + expand_omp_for_static_chunk. */ > + > +void > +f4 (int n) > +{ > + int i; > +#pragma omp teams > +#pragma omp distribute dist_schedule(static, 4) > + for (i = 0; i < n; ++i) > + bar (i); > +} > diff --git a/gcc/testsuite/c-c++-common/gomp/masked-1.c > b/gcc/testsuite/c-c++-common/gomp/masked-1.c > index 8de87b551e3..28d2735d96e 100644 > --- a/gcc/testsuite/c-c++-common/gomp/masked-1.c > +++ b/gcc/testsuite/c-c++-common/gomp/masked-1.c > @@ -26,4 +26,5 @@ foo (int x, int *a) > } > > /* { dg-final { scan-tree-dump "GOMP_has_masked_thread_num" "omplower" } } */ > +/* { dg-final { scan-tree-dump-not "GOMP_masked_end" "omplower" } } */ > /* { dg-final { scan-tree-dump-not "omp_get_thread_num" "omplower" } } */ > diff --git a/gcc/testsuite/c-c++-common/gomp/masked-3.c > b/gcc/testsuite/c-c++-common/gomp/masked-3.c > new file mode 100644 > index 00000000000..adea59fdb69 > --- /dev/null > +++ b/gcc/testsuite/c-c++-common/gomp/masked-3.c > @@ -0,0 +1,15 @@ > +/* { dg-do compile } */ > +/* { dg-additional-options "-fopenmp-ompt -fdump-tree-omplower" } */ > + > +void bar (void); > + > +void > +foo (void) > +{ > + #pragma omp masked > + bar (); > +} > + > +/* { dg-final { scan-tree-dump-times "GOMP_has_masked_thread_num" 1 > "omplower" } } */ > +/* { dg-final { scan-tree-dump-times "GOMP_masked_end" 1 "omplower" } } */ > +/* { dg-final { scan-tree-dump-not "omp_get_thread_num" "omplower" } } */ > diff --git a/gcc/testsuite/c-c++-common/gomp/openmp-ompt-1.c > b/gcc/testsuite/c-c++-common/gomp/openmp-ompt-1.c > new file mode 100644 > index 00000000000..0a45c2bd52c > --- /dev/null > +++ b/gcc/testsuite/c-c++-common/gomp/openmp-ompt-1.c > @@ -0,0 +1,4 @@ > +/* { dg-do compile } */ > +/* { dg-options "-fopenmp-ompt" } */ > + > +/* { dg-error ".-fopenmp-ompt. and .-fopenmp-ompt-detailed. require > .-fopenmp." "" { target *-*-* } 0 } */ > diff --git a/gcc/testsuite/c-c++-common/gomp/openmp-ompt-2.c > b/gcc/testsuite/c-c++-common/gomp/openmp-ompt-2.c > new file mode 100644 > index 00000000000..f93f6920327 > --- /dev/null > +++ b/gcc/testsuite/c-c++-common/gomp/openmp-ompt-2.c > @@ -0,0 +1,4 @@ > +/* { dg-do compile } */ > +/* { dg-options "-fopenmp-ompt-detailed" } */ > + > +/* { dg-error ".-fopenmp-ompt. and .-fopenmp-ompt-detailed. require > .-fopenmp." "" { target *-*-* } 0 } */ > diff --git a/gcc/testsuite/c-c++-common/gomp/single2.c > b/gcc/testsuite/c-c++-common/gomp/single2.c > new file mode 100644 > index 00000000000..215f0368ce4 > --- /dev/null > +++ b/gcc/testsuite/c-c++-common/gomp/single2.c > @@ -0,0 +1,13 @@ > +/* { dg-do compile } */ > +/* { dg-additional-options "-fopenmp-ompt -fdump-tree-ompexp" } */ > + > +void > +foo (void) > +{ > + #pragma omp single > + ; > +} > + > +/* { dg-final { scan-tree-dump-times "GOMP_single_start_with_end" 1 "ompexp" > } } */ > +/* { dg-final { scan-tree-dump-times "GOMP_single_end" 1 "ompexp" } } */ > +/* { dg-final { scan-tree-dump-not "GOMP_single_start \\(" "ompexp" } } */ > diff --git a/libgomp/config/gcn/teams.c b/libgomp/config/gcn/teams.c > index 5aa8b15502a..99f807ac876 100644 > --- a/libgomp/config/gcn/teams.c > +++ b/libgomp/config/gcn/teams.c > @@ -50,6 +50,29 @@ GOMP_distribute_static_worksharing (void) > return nteams + tid * 1I; > } > > +/* OMPT variant enabled by -fopenmp-ompt. */ > + > +_Complex int > +GOMP_distribute_static_worksharing_start (void) > +{ > + int __lds *gomp_team_num = (int __lds *) GOMP_TEAM_NUM; > + unsigned tid = *gomp_team_num; > + unsigned nteams = gomp_num_teams_var + 1; > + return nteams + tid * 1I; > +} > + > +/* Stub for OMPT callback enabled by -fopenmp-ompt-detailed. */ > + > +void > +GOMP_distribute_static_worksharing_dispatch (void) > +{} > + > +/* Stub for OMPT callback enabled by -fopenmp-ompt. */ > + > +void > +GOMP_distribute_static_worksharing_end (void) > +{} > + > int > omp_get_num_teams (void) > { > diff --git a/libgomp/config/nvptx/teams.c b/libgomp/config/nvptx/teams.c > index 9763f6f4108..3e6ea287276 100644 > --- a/libgomp/config/nvptx/teams.c > +++ b/libgomp/config/nvptx/teams.c > @@ -52,6 +52,27 @@ GOMP_distribute_static_worksharing (void) > return nteams + tid * 1I; > } > > +/* OMPT variant enabled by -fopenmp-ompt. */ > + > +_Complex int > +GOMP_distribute_static_worksharing_start (void) > +{ > + unsigned tid = __gomp_team_num; > + unsigned nteams = gomp_num_teams_var + 1; > + return nteams + tid * 1I; > +} > + > +/* Stub for OMPT callback enabled by -fopenmp-ompt-detailed. */ > + > +void > +GOMP_distribute_static_worksharing_dispatch (void) > +{} > + > +/* Stub for OMPT callback enabled by -fopenmp-ompt. */ > + > +void > +GOMP_distribute_static_worksharing_end (void) > +{} > int > omp_get_num_teams (void) > { > diff --git a/libgomp/libgomp.map b/libgomp/libgomp.map > index e2985347bf7..3f465cc1491 100644 > --- a/libgomp/libgomp.map > +++ b/libgomp/libgomp.map > @@ -485,12 +485,21 @@ GOMP_6.0.1 { > GOMP_6.0.2 { > global: > GOMP_has_masked_thread_num; > + GOMP_masked_end; > GOMP_loop_static_worksharing; > + GOMP_loop_static_worksharing_start; > + GOMP_loop_static_worksharing_dispatch; > + GOMP_loop_static_worksharing_end; > GOMP_distribute_static_worksharing; > + GOMP_distribute_static_worksharing_start; > + GOMP_distribute_static_worksharing_dispatch; > + GOMP_distribute_static_worksharing_end; > GOMP_barrier_ext; > GOMP_barrier_cancel_ext; > GOMP_reduction_start; > GOMP_reduction_end; > + GOMP_single_start_with_end; > + GOMP_single_end; > } GOMP_6.0.1; > > OACC_2.0 { > diff --git a/libgomp/libgomp_g.h b/libgomp/libgomp_g.h > index a3097d5b093..4a3565684cc 100644 > --- a/libgomp/libgomp_g.h > +++ b/libgomp/libgomp_g.h > @@ -144,6 +144,11 @@ extern void GOMP_loop_end (void); > extern void GOMP_loop_end_nowait (void); > extern bool GOMP_loop_end_cancel (void); > > +extern __complex__ int GOMP_loop_static_worksharing (void); > +extern __complex__ int GOMP_loop_static_worksharing_start (void); > +extern void GOMP_loop_static_worksharing_dispatch (void); > +extern void GOMP_loop_static_worksharing_end (void); > + > /* loop_ull.c */ > > extern bool GOMP_loop_ull_static_start (bool, unsigned long long, > @@ -298,7 +303,7 @@ extern unsigned GOMP_parallel_reductions (void (*) (void > *), void *, unsigned, > extern bool GOMP_cancel (int, bool); > extern bool GOMP_cancellation_point (int); > extern bool GOMP_has_masked_thread_num (int); > -extern __complex__ int GOMP_loop_static_worksharing (void); > +extern void GOMP_masked_end (void); > > /* task.c */ > > @@ -339,6 +344,8 @@ extern bool GOMP_sections_end_cancel (void); > /* single.c */ > > extern bool GOMP_single_start (void); > +extern bool GOMP_single_start_with_end (void); > +extern void GOMP_single_end (void); > extern void *GOMP_single_copy_start (void); > extern void GOMP_single_copy_end (void *); > > @@ -380,6 +387,9 @@ extern void *GOMP_target_map_indirect_ptr (void *); > extern void GOMP_teams_reg (void (*) (void *), void *, unsigned, unsigned, > unsigned); > extern __complex__ int GOMP_distribute_static_worksharing (void); > +extern __complex__ int GOMP_distribute_static_worksharing_start (void); > +extern void GOMP_distribute_static_worksharing_dispatch (void); > +extern void GOMP_distribute_static_worksharing_end (void); > > /* allocator.c */ > > diff --git a/libgomp/loop.c b/libgomp/loop.c > index e3fbeb1c4de..f286ce6b6db 100644 > --- a/libgomp/loop.c > +++ b/libgomp/loop.c > @@ -1186,3 +1186,38 @@ GOMP_loop_ordered_guided_next (long *istart, long > *iend) > return gomp_loop_ordered_guided_next (istart, iend); > } > #endif > + > +/* For a worksharing-loop construct with static schedule, return the thread > ID > + and number of threads packed into a single complex value. */ > + > +_Complex int > +GOMP_loop_static_worksharing (void) > +{ > + struct gomp_team *team = gomp_thread ()->ts.team; > + unsigned tid = gomp_thread ()->ts.team_id; > + unsigned nthreads = team ? team->nthreads : 1; > + return nthreads + tid * 1I; > +} > + > +/* OMPT variant enabled by -fopenmp-ompt. */ > + > +_Complex int > +GOMP_loop_static_worksharing_start (void) > +{ > + struct gomp_team *team = gomp_thread ()->ts.team; > + unsigned tid = gomp_thread ()->ts.team_id; > + unsigned nthreads = team ? team->nthreads : 1; > + return nthreads + tid * 1I; > +} > + > +/* Stub for OMPT callback enabled by -fopenmp-ompt-detailed. */ > + > +void > +GOMP_loop_static_worksharing_dispatch (void) > +{} > + > +/* Stub for OMPT callback enabled by -fopenmp-ompt. */ > + > +void > +GOMP_loop_static_worksharing_end (void) > +{} > diff --git a/libgomp/parallel.c b/libgomp/parallel.c > index 92472ecb4f0..bb522b39ecb 100644 > --- a/libgomp/parallel.c > +++ b/libgomp/parallel.c > @@ -271,18 +271,6 @@ GOMP_cancel (int which, bool do_cancel) > return true; > } > > -/* For a worksharing-loop construct with static schedule, return the thread > ID > - and number of threads packed into a single complex value. */ > - > -_Complex int > -GOMP_loop_static_worksharing (void) > -{ > - struct gomp_team *team = gomp_thread ()->ts.team; > - unsigned tid = gomp_thread ()->ts.team_id; > - unsigned nthreads = team ? team->nthreads : 1; > - return nthreads + tid * 1I; > -} > - > /* Return true if the current thread number equals TID. > Used to implement the masked construct's filter clause. */ > > @@ -291,6 +279,12 @@ GOMP_has_masked_thread_num (int tid) > { > return tid == gomp_thread ()->ts.team_id; > } > + > +/* Stub for OMPT callback enabled by -fopenmp-ompt. */ > + > +void > +GOMP_masked_end (void) > +{} > > /* The public OpenMP API for thread and team related inquiries. */ > > diff --git a/libgomp/single.c b/libgomp/single.c > index 65126000ea9..00a84e1c9c1 100644 > --- a/libgomp/single.c > +++ b/libgomp/single.c > @@ -55,6 +55,37 @@ GOMP_single_start (void) > #endif > } > > +/* OMPT variant enabled by -fopenmp-ompt. */ > + > +bool > +GOMP_single_start_with_end (void) > +{ > +#ifdef HAVE_SYNC_BUILTINS > + struct gomp_thread *thr = gomp_thread (); > + struct gomp_team *team = thr->ts.team; > + unsigned long single_count; > + > + if (__builtin_expect (team == NULL, 0)) > + return true; > + > + single_count = thr->ts.single_count++; > + return __sync_bool_compare_and_swap (&team->single_count, single_count, > + single_count + 1L); > +#else > + bool ret = gomp_work_share_start (0); > + if (ret) > + gomp_work_share_init_done (); > + gomp_work_share_end_nowait (); > + return ret; > +#endif > +} > + > +/* Stub for OMPT callback enabled by -fopenmp-ompt. */ > + > +void > +GOMP_single_end (void) > +{} > + > /* This routine is called when first encountering a SINGLE construct that > does have a COPYPRIVATE clause. Returns NULL if this is the thread > that should execute the clause; otherwise the return value is pointer > @@ -69,7 +100,7 @@ GOMP_single_copy_start (void) > void *ret; > > first = gomp_work_share_start (0); > - > + > if (first) > { > gomp_work_share_init_done (); > diff --git a/libgomp/teams.c b/libgomp/teams.c > index 7038623b2df..f501123d647 100644 > --- a/libgomp/teams.c > +++ b/libgomp/teams.c > @@ -70,6 +70,29 @@ GOMP_distribute_static_worksharing (void) > return nteams + tid * 1I; > } > > +/* OMPT variant enabled by -fopenmp-ompt. */ > + > +_Complex int > +GOMP_distribute_static_worksharing_start (void) > +{ > + struct gomp_thread *thr = gomp_thread (); > + unsigned tid = thr->team_num; > + unsigned nteams = thr->num_teams + 1; > + return nteams + tid * 1I; > +} > + > +/* Stub for OMPT callback enabled by -fopenmp-ompt-detailed. */ > + > +void > +GOMP_distribute_static_worksharing_dispatch (void) > +{} > + > +/* Stub for OMPT callback enabled by -fopenmp-ompt. */ > + > +void > +GOMP_distribute_static_worksharing_end (void) > +{} > + > int > omp_get_num_teams (void) > { > diff --git a/libgomp/testsuite/libgomp.c-c++-common/for-static-1.c > b/libgomp/testsuite/libgomp.c-c++-common/for-static-1.c > new file mode 100644 > index 00000000000..b304c8c10ec > --- /dev/null > +++ b/libgomp/testsuite/libgomp.c-c++-common/for-static-1.c > @@ -0,0 +1,13 @@ > +/* { dg-do link { target offload_target_any } } */ > +/* { dg-additional-options "-O0 -foffload=-fdump-tree-optimized" } */ > + > +/* Check that the static distribute expanders fetch the team id and count > + through a single GOMP_distribute_static_worksharing call rather than > separate > + omp_get_team_num and omp_get_num_teams calls, and that this function > exists > + on the target side. */ > + > +#include "for-static.h" > + > +/* { dg-final { scan-offload-tree-dump-times > "GOMP_distribute_static_worksharing \\(" 2 "optimized" } } */ > +/* { dg-final { scan-offload-tree-dump-not "omp_get_num_teams" "optimized" } > } */ > +/* { dg-final { scan-offload-tree-dump-not "omp_get_team_num" "optimized" } > } */ > diff --git a/libgomp/testsuite/libgomp.c-c++-common/for-static-2.c > b/libgomp/testsuite/libgomp.c-c++-common/for-static-2.c > new file mode 100644 > index 00000000000..83e577a3298 > --- /dev/null > +++ b/libgomp/testsuite/libgomp.c-c++-common/for-static-2.c > @@ -0,0 +1,11 @@ > +/* { dg-do link { target offload_target_any } } */ > +/* { dg-additional-options "-O0 -foffload=-fdump-tree-optimized > -fopenmp-ompt" } */ > + > +/* Check that, with -fopenmp-ompt, the _start and _end variants are called. > */ > + > +#include "for-static.h" > + > +/* { dg-final { scan-offload-tree-dump-times > "GOMP_distribute_static_worksharing_start \\(" 2 "optimized" } } */ > +/* { dg-final { scan-offload-tree-dump-times > "GOMP_distribute_static_worksharing_end \\(" 2 "optimized" } } */ > +/* { dg-final { scan-offload-tree-dump-not > "GOMP_distribute_static_worksharing \\(" "optimized" } } */ > +/* { dg-final { scan-offload-tree-dump-not > "GOMP_distribute_static_worksharing_dispatch" "optimized" } } */ > diff --git a/libgomp/testsuite/libgomp.c-c++-common/for-static-3.c > b/libgomp/testsuite/libgomp.c-c++-common/for-static-3.c > new file mode 100644 > index 00000000000..e7b904eedc3 > --- /dev/null > +++ b/libgomp/testsuite/libgomp.c-c++-common/for-static-3.c > @@ -0,0 +1,11 @@ > +/* { dg-do link { target offload_target_any } } */ > +/* { dg-additional-options "-O0 -foffload=-fdump-tree-optimized > -fopenmp-ompt-detailed" } */ > + > +/* Check that, with -fopenmp-ompt-detailed, _dispatch builtins are called > along > + with the _start and _end variants. */ > + > +#include "for-static.h" > + > +/* { dg-final { scan-offload-tree-dump-times > "GOMP_distribute_static_worksharing_start \\(" 2 "optimized" } } */ > +/* { dg-final { scan-offload-tree-dump-times > "GOMP_distribute_static_worksharing_dispatch \\(" 2 "optimized" } } */ > +/* { dg-final { scan-offload-tree-dump-times > "GOMP_distribute_static_worksharing_end \\(" 2 "optimized" } } */ > diff --git a/libgomp/testsuite/libgomp.c-c++-common/for-static.h > b/libgomp/testsuite/libgomp.c-c++-common/for-static.h > new file mode 100644 > index 00000000000..a6c0e055b83 > --- /dev/null > +++ b/libgomp/testsuite/libgomp.c-c++-common/for-static.h > @@ -0,0 +1,37 @@ > +void bar (int a) > +{} > + > +/* Distribute without a chunk size goes through > + expand_omp_for_static_nochunk. */ > + > +void > +f3 (int n) > +{ > + int i; > +#pragma omp target > +#pragma omp teams > +#pragma omp distribute > + for (i = 0; i < n; ++i) > + bar (i); > +} > + > +/* Distribute with a chunk size goes through > + expand_omp_for_static_chunk. */ > + > +void > +f4 (int n) > +{ > + int i; > +#pragma omp target > +#pragma omp teams > +#pragma omp distribute dist_schedule(static, 4) > + for (i = 0; i < n; ++i) > + bar (i); > +} > + > +int main (void) > +{ > + f3(0); > + f4(1); > + return 0; > +} > -- > 2.53.0
