On Fri, Oct 17, 2025 at 5:40 PM Josef Melcr <[email protected]> wrote: > > Hi, > I pushed the patched below to master. It is largely the same as the v5 > (link: https://gcc.gnu.org/pipermail/gcc-patches/2025-October/697470.html), > but I took out all the oacc bits, as they did not apply (discussed in > https://gcc.gnu.org/pipermail/gcc-patches/2025-October/697527.html). > > ========== > > This patch enables constant propagation to outlined OpenMP kernels. > It does so using a new function attribute called ' callback' (note the > space). > > The attribute ' callback' captures the notion of a function calling one > of its arguments with some of its parameters as arguments. An OpenMP > example of such function is GOMP_parallel. > We implement the attribute with new callgraph edges called callback > edges. They are imaginary edges pointing from the caller of the function > with the attribute (e.g. caller of GOMP_parallel) to the body function > itself (e.g. the outlined OpenMP body). They share their call statement > with the edge from which they are derived (direct edge caller -> GOMP_parallel > in this case). These edges allow passes such as ipa-cp to see the hidden > call site to the body function and optimize the function accordingly. > > To illustrate on an example, the body GOMP_parallel looks something > like this: > > void GOMP_parallel (void (*fn) (void *), void *data, /* ... */) > { > /* ... */ > fn (data); > /* ... */ > } > > > If we extend it with the attribute ' callback(1, 2)', we express that the > function calls its first argument and passes it its second argument. > This is represented in the call graph in this manner: > > direct indirect > caller -----------------> GOMP_parallel ---------------> fn > | > ----------------------> fn > callback > > The direct edge is then the callback-carrying edge, all new edges > are the derived callback edges. > While constant propagation is the main focus of this patch, callback > edges can be useful for different passes (for example, they improve icf > for OpenMP kernels), as they allow for address redirection. > If the outlined body function gets optimized and cloned, from body_fn to > body_fn.optimized, the callback edge allows us to replace the > address in the arguments list: > > GOMP_parallel (body_fn, &data_struct, /* ... */); > > becomes > > GOMP_parallel (body_fn.optimized, &data_struct, /* ... */); > > This redirection is possible for any function with the attribute. > > This callback attribute implementation is partially compatible with > clang's implementation. Its semantics, arguments and argument indexing style > are > the same, but we represent an unknown argument position with 0 > (precedent set by attributes such as 'format'), while clang uses -1 or '?'. > We use the index 1 for the 'this' pointer in member functions, clang > uses 0. We also allow for multiple callback attributes on the same function, > while clang only allows one. > > The attribute is currently for GCC internal use only, thanks to the > space in its name. Originally, it was supposed to be called > 'callback' like its clang counterpart, but we cannot use this name, as > clang uses non-standard indexing style, leading to inconsistencies. The > attribute will be introduced into the public API as 'gnu::callback_only' > in a future patch. > > The attribute allows us to propagate constants into body functions of > OpenMP constructs. Currently, GCC won't propagate the value 'c' into the > OpenMP body in the following example: > > int a[100]; > void test(int c) { > #pragma omp parallel for > for (int i = 0; i < c; i++) { > if (!__builtin_constant_p(c)) { > __builtin_abort(); > } > a[i] = i; > } > } > int main() { > test(100); > return a[5] - 5; > } > > With this patch, the body function will get cloned and the constant 'c' > will get propagated. > > Some functions may utilize the attribute's infrastructure without being > declared with it, for example GOMP_task. These functions are special > cases and use the special case functions found in attr-callback.h. Special > cases use the attribute under certain circumstances, for example > GOMP_task uses it when the copy function is not being used required. > > gcc/ChangeLog: > > * Makefile.in: Add attr-callback.o to OBJS. > * builtin-attrs.def (ATTR_CALLBACK): Callback attr identifier. > (DEF_CALLBACK_ATTRIBUTE): Macro for callback attr creation. > (GOMP): Attr for libgomp functions. > (ATTR_CALLBACK_GOMP_LIST): ATTR_NOTHROW_LIST with GOMP callback > attr added. > * cgraph.cc (cgraph_add_edge_to_call_site_hash): Always hash the > callback-carrying edge. > (cgraph_node::get_edge): Always return the callback-carrying > edge. > (cgraph_edge::set_call_stmt): Add cascade for callback edges. > (symbol_table::create_edge): Allow callback edges to share call > stmts, initialize new flags. > (cgraph_edge::make_callback): New method, derives a new callback > edge. > (cgraph_edge::get_callback_carrying_edge): New method. > (cgraph_edge::first_callback_edge): Likewise. > (cgraph_edge::next_callback_edge): Likewise. > (cgraph_edge::purge_callback_edges): Likewise. > (cgraph_edge::redirect_callee): When redirecting a callback > edge, redirect its ref as well. > (cgraph_edge::redirect_call_stmt_to_callee): Add callback edge > redirection logic, set update_derived_edges to true hwne > redirecting the carrying edge. > (cgraph_node::remove_callers): Add cascade for callback edges. > (cgraph_edge::dump_edge_flags): Print callback flags. > (cgraph_node::verify_node): Add sanity checks for callback > edges. > * cgraph.h: Add new 1 bit flags and 16 bit callback_id to > cgraph_edge class. > * cgraphclones.cc (cgraph_edge::clone): Copy over callback data. > * cif-code.def (CALLBACK_EDGE): Add CIF_CALLBACK_EDGE code. > * ipa-cp.cc (purge_useless_callback_edges): New function, > deletes callback edges when necessary. > (ipcp_decision_stage): Call purge_useless_callback_edges. > * ipa-fnsummary.cc (ipa_call_summary_t::duplicate): Add > an exception for callback edges. > (analyze_function_body): Copy over summary from carrying to > callback edge. > * ipa-inline-analysis.cc (do_estimate_growth_1): Skip callback > edges when estimating growth. > * ipa-inline-transform.cc (inline_transform): Add redirection > cascade for callback edges. > * ipa-param-manipulation.cc > (drop_decl_attribute_if_params_changed_p): New function. > (ipa_param_adjustments::build_new_function_type): Add > args_modified out param. > (ipa_param_adjustments::adjust_decl): Drop callback attrs when > modifying args. > * ipa-param-manipulation.h: Adjust decl of > build_new_function_type. > * ipa-prop.cc (ipa_duplicate_jump_function): Add decl. > (init_callback_edge_summary): New function. > (ipa_compute_jump_functions_for_edge): Add callback edge > creation logic. > * lto-cgraph.cc (lto_output_edge): Stream out callback data. > (input_edge): Input callback data. > * omp-builtins.def (BUILT_IN_GOMP_PARALLEL_LOOP_STATIC): Use new > attr list. > (BUILT_IN_GOMP_PARALLEL_LOOP_GUIDED): Likewise. > (BUILT_IN_GOMP_PARALLEL_LOOP_NONMONOTONIC_DYNAMIC): Likewise. > (BUILT_IN_GOMP_PARALLEL_LOOP_NONMONOTONIC_RUNTIME): Likewise. > (BUILT_IN_GOMP_PARALLEL): Likewise. > (BUILT_IN_GOMP_PARALLEL_SECTIONS): Likewise. > (BUILT_IN_GOMP_TEAMS_REG): Likewise. > * tree-core.h (ECF_CB_1_2): New constant for callback(1,2). > * tree-inline.cc (copy_bb): Copy callback edges when copying the > carrying edge. > (redirect_all_calls): Redirect callback edges. > * tree.cc (set_call_expr_flags): Create callback attr according > to the ECF_CB flag. > * attr-callback.cc: New file. > * attr-callback.h: New file. > > gcc/c-family/ChangeLog: > > * c-attribs.cc: Define callback attr. > > gcc/fortran/ChangeLog: > > * f95-lang.cc (ATTR_CALLBACK_GOMP_LIST): New attr list > corresponding to the list in builtin-attrs.def. > > gcc/testsuite/ChangeLog: > > * gcc.dg/ipa/ipcp-cb-spec1.c: New test. > * gcc.dg/ipa/ipcp-cb-spec2.c: New test. > * gcc.dg/ipa/ipcp-cb1.c: New test.
These run-time tests should be moved to libgomp to avoid xgcc: fatal error: cannot read spec file 'libgomp.spec': No such file or directory compilation terminated. compiler exited with status 1 FAIL: gcc.dg/ipa/ipcp-cb1.c (test for excess errors) -- H.J.
