Hi! On Tue, 31 May 2016 15:35:17 -0700, Cesar Philippidis <ce...@codesourcery.com> wrote: > On 05/31/2016 08:45 AM, Thomas Schwinge wrote: > > > While working on something else, I came across the following. Cesar, can > > you please verify that this is really dead code in the Fortran front end, > > which currently is the only producer of OMP_CLAUSE_DEVICE_RESIDENT? > > You're correct. Declare should be using map clauses for device resident > and link.
> > Also, I noticed that the Fortran front end never creates OMP_CLAUSE_MAP > > with GOMP_MAP_LINK or GOMP_MAP_DEVICE_RESIDENT -- how is OpenACC declare > > working at all? Cesar, if this is not a simple question to answer, > > please file a GCC PR, for somebody to have a look later. > > Declare might not be working in fortran, or at least not all of its > clauses. I'll add this on my todo list. I extended the scope of the existing <https://gcc.gnu.org/PR63859> to generally examine the "Fortran OpenACC declare directive". Committed in r236985: commit 9c9a6c2515033aabbba5a037674573526ab37319 Author: tschwinge <tschwinge@138bc75d-0d04-0410-961f-82ee72b054a4> Date: Wed Jun 1 11:55:35 2016 +0000 Remove the unused OMP_CLAUSE_DEVICE_RESIDENT gcc/ * tree-core.h (enum omp_clause_code): Remove OMP_CLAUSE_DEVICE_RESIDENT. Adjust all users. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@236985 138bc75d-0d04-0410-961f-82ee72b054a4 --- gcc/ChangeLog | 5 +++++ gcc/fortran/trans-openmp.c | 3 --- gcc/gimplify.c | 5 ----- gcc/omp-low.c | 2 -- gcc/tree-core.h | 3 --- gcc/tree-pretty-print.c | 3 --- gcc/tree.c | 3 --- 7 files changed, 5 insertions(+), 19 deletions(-) diff --git gcc/ChangeLog gcc/ChangeLog index 0bb7f28..58fea2f 100644 --- gcc/ChangeLog +++ gcc/ChangeLog @@ -1,3 +1,8 @@ +2016-06-01 Thomas Schwinge <tho...@codesourcery.com> + + * tree-core.h (enum omp_clause_code): Remove + OMP_CLAUSE_DEVICE_RESIDENT. Adjust all users. + 2016-06-01 Kyrylo Tkachov <kyrylo.tkac...@arm.com> * config/arm/sync.md (arm_store_exclusive<mode>): diff --git gcc/fortran/trans-openmp.c gcc/fortran/trans-openmp.c index c2d89eb..d3276f9 100644 --- gcc/fortran/trans-openmp.c +++ gcc/fortran/trans-openmp.c @@ -1773,9 +1773,6 @@ gfc_trans_omp_clauses (stmtblock_t *block, gfc_omp_clauses *clauses, case OMP_LIST_USE_DEVICE: clause_code = OMP_CLAUSE_USE_DEVICE_PTR; goto add_clause; - case OMP_LIST_DEVICE_RESIDENT: - clause_code = OMP_CLAUSE_DEVICE_RESIDENT; - goto add_clause; add_clause: omp_clauses diff --git gcc/gimplify.c gcc/gimplify.c index 8b7dddc..131fa24 100644 --- gcc/gimplify.c +++ gcc/gimplify.c @@ -7538,10 +7538,6 @@ gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p, } break; - case OMP_CLAUSE_DEVICE_RESIDENT: - remove = true; - break; - case OMP_CLAUSE_NOWAIT: case OMP_CLAUSE_ORDERED: case OMP_CLAUSE_UNTIED: @@ -8275,7 +8271,6 @@ gimplify_adjust_omp_clauses (gimple_seq *pre_p, gimple_seq body, tree *list_p, case OMP_CLAUSE__CILK_FOR_COUNT_: case OMP_CLAUSE_ASYNC: case OMP_CLAUSE_WAIT: - case OMP_CLAUSE_DEVICE_RESIDENT: case OMP_CLAUSE_INDEPENDENT: case OMP_CLAUSE_NUM_GANGS: case OMP_CLAUSE_NUM_WORKERS: diff --git gcc/omp-low.c gcc/omp-low.c index a11f44b..77bdb18 100644 --- gcc/omp-low.c +++ gcc/omp-low.c @@ -2200,7 +2200,6 @@ scan_sharing_clauses (tree clauses, omp_context *ctx, install_var_local (decl, ctx); break; - case OMP_CLAUSE_DEVICE_RESIDENT: case OMP_CLAUSE__CACHE_: sorry ("Clause not supported yet"); break; @@ -2368,7 +2367,6 @@ scan_sharing_clauses (tree clauses, omp_context *ctx, case OMP_CLAUSE__GRIDDIM_: break; - case OMP_CLAUSE_DEVICE_RESIDENT: case OMP_CLAUSE__CACHE_: sorry ("Clause not supported yet"); break; diff --git gcc/tree-core.h gcc/tree-core.h index b069928..db5b470 100644 --- gcc/tree-core.h +++ gcc/tree-core.h @@ -316,9 +316,6 @@ enum omp_clause_code { #pragma acc cache (variable-list). */ OMP_CLAUSE__CACHE_, - /* OpenACC clause: device_resident (variable_list). */ - OMP_CLAUSE_DEVICE_RESIDENT, - /* OpenACC clause: gang [(gang-argument-list)]. Where gang-argument-list: [gang-argument-list, ] gang-argument diff --git gcc/tree-pretty-print.c gcc/tree-pretty-print.c index 0e7fdd1..734ecda 100644 --- gcc/tree-pretty-print.c +++ gcc/tree-pretty-print.c @@ -407,9 +407,6 @@ dump_omp_clause (pretty_printer *pp, tree clause, int spc, int flags) case OMP_CLAUSE__LOOPTEMP_: name = "_looptemp_"; goto print_remap; - case OMP_CLAUSE_DEVICE_RESIDENT: - name = "device_resident"; - goto print_remap; case OMP_CLAUSE_TO_DECLARE: name = "to"; goto print_remap; diff --git gcc/tree.c gcc/tree.c index f4b470b..7511d0a 100644 --- gcc/tree.c +++ gcc/tree.c @@ -281,7 +281,6 @@ unsigned const char omp_clause_num_ops[] = 1, /* OMP_CLAUSE_USE_DEVICE_PTR */ 1, /* OMP_CLAUSE_IS_DEVICE_PTR */ 2, /* OMP_CLAUSE__CACHE_ */ - 1, /* OMP_CLAUSE_DEVICE_RESIDENT */ 2, /* OMP_CLAUSE_GANG */ 1, /* OMP_CLAUSE_ASYNC */ 1, /* OMP_CLAUSE_WAIT */ @@ -353,7 +352,6 @@ const char * const omp_clause_code_name[] = "use_device_ptr", "is_device_ptr", "_cache_", - "device_resident", "gang", "async", "wait", @@ -11764,7 +11762,6 @@ walk_tree_1 (tree *tp, walk_tree_fn func, void *data, WALK_SUBTREE (OMP_CLAUSE_OPERAND (*tp, 1)); /* FALLTHRU */ - case OMP_CLAUSE_DEVICE_RESIDENT: case OMP_CLAUSE_ASYNC: case OMP_CLAUSE_WAIT: case OMP_CLAUSE_WORKER: Grüße Thomas
