Hi Jakub! When adding support for C++ templates usage with OpenACC directives (gcc/cp/pt.c:tsubst_expr), we found:
On Mon, 27 Apr 2015 10:57:14 -0700, Cesar Philippidis
<cesar_philippi...@mentor.com> wrote:
> On 04/27/2015 10:40 AM, Thomas Schwinge wrote:
> > Cesar wrote:
> >> case OMP_TARGET_UPDATE:
> >> tmp = tsubst_omp_clauses (OMP_TARGET_UPDATE_CLAUSES (t), false,
> >> args, complain, in_decl);
> >> @@ -14253,6 +14292,16 @@ tsubst_expr (tree t, tree args, tsubst_flags_t
> >> complain, tree in_decl,
> >> add_stmt (t);
> >> break;
> >>
> >> + case OACC_ENTER_DATA:
> >> + case OACC_EXIT_DATA:
> >> + case OACC_UPDATE:
> >> + tmp = tsubst_omp_clauses (TREE_OPERAND (t, 0), false,
> >> + args, complain, in_decl);
> >> + t = copy_node (t);
> >> + TREE_OPERAND (t, 0) = tmp;
> >> + add_stmt (t);
> >> + break;
> >
> > Given the handling for similar codes, why not handle those together with
> > OMP_TARGET_UPDATE, replacing OMP_TARGET_UPDATE_CLAUSES with plain
> > OMP_CLAUSES?
>
> I left it them separate because OMP_CLAUSES expects the clauses to
> appear in tree_operand 1, but oacc_update, oacc_enter_data and
> oacc_exit_data have the clauses at tree_operand 0. I think this may have
> been a bug with openmp, so I was planning on informing Jakub when I
> posted this patch upstream.
>
> The attached patch cleans up [...]
> --- a/gcc/cp/pt.c
> +++ b/gcc/cp/pt.c
> @@ -14277,21 +14277,7 @@ tsubst_expr (tree t, tree args, tsubst_flags_t
> complain, tree in_decl,
> add_stmt (t);
> break;
>
> case OMP_TARGET_UPDATE:
> - tmp = tsubst_omp_clauses (OMP_TARGET_UPDATE_CLAUSES (t), false,
> - args, complain, in_decl);
> - t = copy_node (t);
> - OMP_CLAUSES (t) = tmp;
> - add_stmt (t);
> - break;
> -
> case OACC_ENTER_DATA:
> case OACC_EXIT_DATA:
> case OACC_UPDATE:
I guess nobody so far ;-) has been using OpenMP's target update directive
in templated code -- OK to commit the following, and to which branches
(4.9, 5, trunk)?
commit 5ea85b95c1d0ccb98d73345f105bf76839b16433
Author: Thomas Schwinge <tho...@codesourcery.com>
Date: Tue Apr 28 20:29:26 2015 +0200
Fix OpenMP's target update directive in templated code.
FAIL: g++.dg/gomp/tpl-target-update.C -std=c++98 (internal compiler
error)
FAIL: g++.dg/gomp/tpl-target-update.C -std=c++98 (test for excess
errors)
FAIL: g++.dg/gomp/tpl-target-update.C -std=c++11 (internal compiler
error)
FAIL: g++.dg/gomp/tpl-target-update.C -std=c++11 (test for excess
errors)
FAIL: g++.dg/gomp/tpl-target-update.C -std=c++14 (internal compiler
error)
FAIL: g++.dg/gomp/tpl-target-update.C -std=c++14 (test for excess
errors)
[...]/source-gcc/gcc/testsuite/g++.dg/gomp/tpl-target-update.C: In
instantiation of 'void f(T, T) [with T = int]':
[...]/source-gcc/gcc/testsuite/g++.dg/gomp/tpl-target-update.C:19:9:
required from here
[...]/source-gcc/gcc/testsuite/g++.dg/gomp/tpl-target-update.C:10:9:
internal compiler error: tree check: expected oacc_parallel or oacc_kernels or
oacc_data or oacc_host_data or omp_parallel or omp_task or omp_for or omp_simd
or cilk_simd or cilk_for or omp_distribute or oacc_loop or omp_teams or
omp_target_data or omp_target or omp_sections or omp_single, have
omp_target_update in tsubst_expr, at cp/pt.c:14209
0xf5aae1 tree_range_check_failed(tree_node const*, char const*, int,
char const*, tree_code, tree_code)
[...]/source-gcc/gcc/tree.c:9384
0x66e201 tree_range_check
[...]/source-gcc/gcc/tree.h:2979
0x66e201 tsubst_expr
[...]/source-gcc/gcc/cp/pt.c:14209
0x6695e3 tsubst_expr
[...]/source-gcc/gcc/cp/pt.c:13752
0x66ac07 tsubst_expr
[...]/source-gcc/gcc/cp/pt.c:13938
0x667c41 instantiate_decl(tree_node*, int, bool)
[...]/source-gcc/gcc/cp/pt.c:20367
0x6ae386 instantiate_pending_templates(int)
[...]/source-gcc/gcc/cp/pt.c:20484
0x6edc3d cp_write_global_declarations()
[...]/source-gcc/gcc/cp/decl2.c:4456
gcc/cp/
* pt.c (tsubst_expr) <OMP_TARGET_UPDATE>: Use
OMP_TARGET_UPDATE_CLAUSES instead of OMP_CLAUSES.
gcc/testsuite/
* g++.dg/gomp/tpl-target-update.C: New file.
---
gcc/cp/pt.c | 2 +-
gcc/testsuite/g++.dg/gomp/tpl-target-update.C | 20 ++++++++++++++++++++
2 files changed, 21 insertions(+), 1 deletion(-)
diff --git gcc/cp/pt.c gcc/cp/pt.c
index f9a5c3b..129517a 100644
--- gcc/cp/pt.c
+++ gcc/cp/pt.c
@@ -14206,7 +14206,7 @@ tsubst_expr (tree t, tree args, tsubst_flags_t
complain, tree in_decl,
tmp = tsubst_omp_clauses (OMP_TARGET_UPDATE_CLAUSES (t), false,
args, complain, in_decl);
t = copy_node (t);
- OMP_CLAUSES (t) = tmp;
+ OMP_TARGET_UPDATE_CLAUSES (t) = tmp;
add_stmt (t);
break;
diff --git gcc/testsuite/g++.dg/gomp/tpl-target-update.C
gcc/testsuite/g++.dg/gomp/tpl-target-update.C
new file mode 100644
index 0000000..6226ebf
--- /dev/null
+++ gcc/testsuite/g++.dg/gomp/tpl-target-update.C
@@ -0,0 +1,20 @@
+// { dg-do compile }
+
+template <typename T>
+void f(T A, T B)
+{
+ extern int *v;
+ T a = 2;
+ T b = 4;
+
+#pragma omp target update to(v[a:b])
+ v[a] = 0;
+
+#pragma omp target update to(v[A:B])
+ v[a] = 0;
+}
+
+void g()
+{
+ f(1, 5);
+}
That said, what is the preferred approach to add support for
OACC_ENTER_DATA, OACC_EXIT_DATA, OACC_UPDATE? I'm not sure hard-coding
TREE_OPERAND (t, 0) in gcc/cp/pt.c:tsubst_expr is the way to go, and also
duplicating the OMP_TARGET_UPDATE code for each of the three OACC_*
doesn't sound appealing -- especially given that we just "switch"ed on
the respective tree code, so the O*_CHECK of the respective O*_CLAUSES
macro is completely redundant. Is it OK to extend gcc/tree.h:OMP_CLAUSES
so that it can also be used for tree codes that keep clauses in operand
0, and then use that here (and also in
gcc/gimplify.c:gimplify_omp_target_update, for example)?
Grüße,
Thomas
pgpTwRx5qiliG.pgp
Description: PGP signature
