[gcc r17-2288] openmp: Add OMPT entry points for worksharing-loop, single, and masked constructs
Thomas Schwinge via Gcc-cvs <[email protected]>
| Newsgroups | gmane.comp.gcc.cvs |
|---|---|
| Message-ID | <[email protected]> |
https://gcc.gnu.org/g:ab0dd7d5e16fcd90a58be84ec45a8a46994cf176 commit r17-2288-gab0dd7d5e16fcd90a58be84ec45a8a46994cf176 Author: Paul-Antoine Arras <[email protected]> Date: Tue Jun 30 11:06:44 2026 +0200 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. Diff: --- 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 ------------- gcc/testsuite/c-c++-common/gomp/for-static-1.c | 17 ++++ gcc/testsuite/c-c++-common/gomp/for-static-2.c | 15 +++ gcc/testsuite/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 +++ gcc/testsuite/c-c++-common/gomp/openmp-ompt-1.c | 4 + gcc/testsuite/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 +++++ .../testsuite/libgomp.c-c++-common/for-static-1.c | 13 +++ .../testsuite/libgomp.c-c++-common/for-static-2.c | 11 ++ .../testsuite/libgomp.c-c++-common/for-static-3.c | 11 ++ .../testsuite/libgomp.c-c++-common/for-static.h | 37 +++++++ 29 files changed, 556 insertions(+), 89 deletions(-) diff --git a/gcc/common.opt b/gcc/common.opt index 0055578ccf60..ecebc186af77 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 490146dde8c0..567a8dfc2d6a 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 606d666cd51b..354ceb76f538 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}. @@ -5601,6 +5602,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 d7440a42761b..465aeebefdb7 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 3d799d280329..ad380b27d062 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 ad5b2225d279..c578e8ece49a 100644 --- a/gcc/omp-low.cc +++ b/gcc/omp-low.cc @@ -8825,7 +8825,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); @@ -8962,6 +8964,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); @@ -9138,6 +9149,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 5522f41f9069..0d96fb6d78e5 100644 --- a/gcc/opts.cc +++ b/gcc/opts.cc @@ -1537,6 +1537,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 a894aaaaed11..000000000000 --- 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 000000000000..50d4e176f3d8 --- /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 000000000000..c3d25b97bed7 --- /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 000000000000..6fd1da6c4075 --- /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 000000000000..6504c75a13ed --- /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 8de87b551e30..28d2735d96e3 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 000000000000..adea59fdb69e --- /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 000000000000..0a45c2bd52c9 --- /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 000000000000..f93f69203274 --- /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 000000000000..215f0368ce4f --- /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 5aa8b15502aa..99f807ac8766 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 9763f6f41084..3e6ea2872766 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 e2985347bf77..3f465cc14916 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 a3097d5b0936..4a3565684cc9 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 e3fbeb1c4de9..f286ce6b6db5 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 92472ecb4f0e..bb522b39ecb1 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 65126000ea9a..00a84e1c9c1c 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 7038623b2dfb..f501123d6479 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 000000000000..b304c8c10ecc --- /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 000000000000..83e577a32980 --- /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 000000000000..e7b904eedc3a --- /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 000000000000..a6c0e055b835 --- /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; +}