public inbox for gcc-cvs@sourceware.org help / color / mirror / Atom feed
From: Jakub Jelinek <jakub@gcc.gnu.org> To: gcc-cvs@gcc.gnu.org Subject: [gcc(refs/vendors/redhat/heads/gcc-10-branch)] Fix OpenMP offload handling for target-link variables for nvptx (PR81689) Date: Wed, 25 Mar 2020 10:12:38 +0000 (GMT) [thread overview] Message-ID: <20200325101238.3C9F8385E00F@sourceware.org> (raw) https://gcc.gnu.org/g:c2211a60ff05b7a0289d3e287e72c181bb4d5d8b commit c2211a60ff05b7a0289d3e287e72c181bb4d5d8b Author: Tobias Burnus <tobias@codesourcery.com> Date: Tue Mar 24 15:13:56 2020 +0100 Fix OpenMP offload handling for target-link variables for nvptx (PR81689) PR libgomp/81689 * lto.c (offload_handle_link_vars): Propagate TREE_PUBLIC state. PR libgomp/81689 * omp-offload.c (omp_finish_file): Fix target-link handling if targetm_common.have_named_sections is false. PR libgomp/81689 * testsuite/libgomp.c/target-link-1.c: Remove xfail. Diff: --- gcc/ChangeLog | 6 ++++++ gcc/lto/ChangeLog | 5 +++++ gcc/lto/lto.c | 1 + gcc/omp-offload.c | 14 +++++++++++++- libgomp/ChangeLog | 5 +++++ libgomp/testsuite/libgomp.c/target-link-1.c | 3 --- 6 files changed, 30 insertions(+), 4 deletions(-) diff --git a/gcc/ChangeLog b/gcc/ChangeLog index 45b534b24ab..ca017bedb1b 100644 --- a/gcc/ChangeLog +++ b/gcc/ChangeLog @@ -1,3 +1,9 @@ +2020-03-24 Tobias Burnus <tobias@codesourcery.com> + + PR libgomp/81689 + * omp-offload.c (omp_finish_file): Fix target-link handling if + targetm_common.have_named_sections is false. + 2020-03-24 Jakub Jelinek <jakub@redhat.com> PR target/94286 diff --git a/gcc/lto/ChangeLog b/gcc/lto/ChangeLog index b3c2138aaa3..619a42d5142 100644 --- a/gcc/lto/ChangeLog +++ b/gcc/lto/ChangeLog @@ -1,3 +1,8 @@ +2020-03-24 Tobias Burnus <tobias@codesourcery.com> + + PR libgomp/81689 + * lto.c (offload_handle_link_vars): Propagate TREE_PUBLIC state. + 2020-01-29 Tobias Burnus <tobias@codesourcery.com> * lto.c (offload_handle_link_vars): Reduce chance of diff --git a/gcc/lto/lto.c b/gcc/lto/lto.c index cd34d6c9e7a..1c37814bde4 100644 --- a/gcc/lto/lto.c +++ b/gcc/lto/lto.c @@ -566,6 +566,7 @@ offload_handle_link_vars (void) "linkptr"), type); TREE_USED (link_ptr_var) = 1; TREE_STATIC (link_ptr_var) = 1; + TREE_PUBLIC (link_ptr_var) = TREE_PUBLIC (var->decl); DECL_ARTIFICIAL (link_ptr_var) = 1; SET_DECL_ASSEMBLER_NAME (link_ptr_var, DECL_NAME (link_ptr_var)); SET_DECL_VALUE_EXPR (var->decl, build_simple_mem_ref (link_ptr_var)); diff --git a/gcc/omp-offload.c b/gcc/omp-offload.c index 11412e1059f..c66f38b6f0c 100644 --- a/gcc/omp-offload.c +++ b/gcc/omp-offload.c @@ -220,7 +220,19 @@ omp_finish_file (void) for (unsigned i = 0; i < num_vars; i++) { tree it = (*offload_vars)[i]; - targetm.record_offload_symbol (it); +#ifdef ACCEL_COMPILER + if (DECL_HAS_VALUE_EXPR_P (it) + && lookup_attribute ("omp declare target link", + DECL_ATTRIBUTES (it))) + { + tree value_expr = DECL_VALUE_EXPR (it); + tree link_ptr_decl = TREE_OPERAND (value_expr, 0); + targetm.record_offload_symbol (link_ptr_decl); + varpool_node::finalize_decl (link_ptr_decl); + } + else +#endif + targetm.record_offload_symbol (it); } } } diff --git a/libgomp/ChangeLog b/libgomp/ChangeLog index c90cbdcc711..7f5a1173eb9 100644 --- a/libgomp/ChangeLog +++ b/libgomp/ChangeLog @@ -1,3 +1,8 @@ +2020-03-24 Tobias Burnus <tobias@codesourcery.com> + + PR libgomp/81689 + * testsuite/libgomp.c/target-link-1.c: Remove xfail. + 2020-03-20 Tobias Burnus <tobias@codesourcery.com> PR libgomp/94251 diff --git a/libgomp/testsuite/libgomp.c/target-link-1.c b/libgomp/testsuite/libgomp.c/target-link-1.c index 99ce33bc9b4..681677cc2aa 100644 --- a/libgomp/testsuite/libgomp.c/target-link-1.c +++ b/libgomp/testsuite/libgomp.c/target-link-1.c @@ -1,6 +1,3 @@ -/* { dg-xfail-if "#pragma omp target link not implemented" { offload_target_nvptx } } - Cf. https://gcc.gnu.org/PR81689. */ - struct S { int s, t; }; int a = 1, b = 1;
reply other threads:[~2020-03-25 10:12 UTC|newest] Thread overview: [no followups] expand[flat|nested] mbox.gz Atom feed
Reply instructions: You may reply publicly to this message via plain-text email using any one of the following methods: * Save the following mbox file, import it into your mail client, and reply-to-all from there: mbox Avoid top-posting and favor interleaved quoting: https://en.wikipedia.org/wiki/Posting_style#Interleaved_style * Reply using the --to, --cc, and --in-reply-to switches of git-send-email(1): git send-email \ --in-reply-to=20200325101238.3C9F8385E00F@sourceware.org \ --to=jakub@gcc.gnu.org \ --cc=gcc-cvs@gcc.gnu.org \ /path/to/YOUR_REPLY https://kernel.org/pub/software/scm/git/docs/git-send-email.html * If your mail client supports setting the In-Reply-To header via mailto: links, try the mailto: linkBe sure your reply has a Subject: header at the top and a blank line before the message body.
This is a public inbox, see mirroring instructions for how to clone and mirror all data and code used for this inbox; as well as URLs for read-only IMAP folder(s) and NNTP newsgroup(s).