Message ID | 87wpmaqlii.fsf@kepler.schwinge.homeip.net |
---|---|
State | New |
Headers | show |
On Tue, May 31, 2016 at 05:45:25PM +0200, 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? > > 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. > > Jakub, if the following patch is OK, should I also clean it up on release > branches as applicable, or just on trunk? It is ok for trunk, not worth backporting to any branches. But your ChangeLog doesn't cover the changes you've actually done. > Remove the unused OMP_CLAUSE_DEVICE_RESIDENT > > gcc/ > * tree-core.h (enum omp_clause_code): Remove > OMP_CLAUSE_DEVICE_RESIDENT. Adjust all users. Jakub
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. I think I ran into this last week when I backported Jakub's recent gfc_match_omp_clauses changes to gomp4. > 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. Right now I working on a subarray problem. Cesar
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 8316bb8..e53a2d3 100644 --- gcc/gimplify.c +++ gcc/gimplify.c @@ -7531,10 +7531,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: @@ -8268,7 +8264,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: