From 1002a7ace111d746249fdea71af9b8e039cea0eb Mon Sep 17 00:00:00 2001 From: Tobias Burnus Date: Wed, 23 Mar 2022 09:44:39 +0100 Subject: LTO: Fixes for renaming issues with offload/OpenMP [PR104285] gcc/lto/ChangeLog: PR middle-end/104285 * lto-partition.cc (maybe_rewrite_identifier): Use get_identifier for the returned string to be usable as hash key. (validize_symbol_for_target): Hence, use return value directly. (privatize_symbol_name_1): Track maybe_rewrite_identifier renames. * lto.cc (offload_handle_link_vars): Move function up before ... (do_whole_program_analysis): Call it after static renamings. (lto_main): Move call after static renamings. libgomp/ChangeLog: PR middle-end/104285 * testsuite/libgomp.c++/target-same-name-2-a.C: New test. * testsuite/libgomp.c++/target-same-name-2-b.C: New test. * testsuite/libgomp.c++/target-same-name-2.C: New test. * testsuite/libgomp.c-c++-common/target-same-name-1-a.c: New test. * testsuite/libgomp.c-c++-common/target-same-name-1-b.c: New test. * testsuite/libgomp.c-c++-common/target-same-name-1.c: New test. --- .../testsuite/libgomp.c++/target-same-name-2-a.C | 50 ++++++++++++++++++ .../testsuite/libgomp.c++/target-same-name-2-b.C | 50 ++++++++++++++++++ libgomp/testsuite/libgomp.c++/target-same-name-2.C | 24 +++++++++ .../libgomp.c-c++-common/target-same-name-1-a.c | 60 ++++++++++++++++++++++ .../libgomp.c-c++-common/target-same-name-1-b.c | 60 ++++++++++++++++++++++ .../libgomp.c-c++-common/target-same-name-1.c | 46 +++++++++++++++++ 6 files changed, 290 insertions(+) create mode 100644 libgomp/testsuite/libgomp.c++/target-same-name-2-a.C create mode 100644 libgomp/testsuite/libgomp.c++/target-same-name-2-b.C create mode 100644 libgomp/testsuite/libgomp.c++/target-same-name-2.C create mode 100644 libgomp/testsuite/libgomp.c-c++-common/target-same-name-1-a.c create mode 100644 libgomp/testsuite/libgomp.c-c++-common/target-same-name-1-b.c create mode 100644 libgomp/testsuite/libgomp.c-c++-common/target-same-name-1.c (limited to 'libgomp') diff --git a/libgomp/testsuite/libgomp.c++/target-same-name-2-a.C b/libgomp/testsuite/libgomp.c++/target-same-name-2-a.C new file mode 100644 index 00000000000..1cff1c8d0c5 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-same-name-2-a.C @@ -0,0 +1,50 @@ +/* { dg-skip-if "" { *-*-* } } */ +/* Used by target-same-name-2.c */ + +#include + +template +int +test_map () +{ + std::complex a(2, 1), a_check; +#pragma omp target map(from : a_check) + { + a_check = a; + } + if (a == a_check) + return 42; + return 0; +} + +template +static int +test_map_static () +{ + std::complex a(-4, 5), a_check; +#pragma omp target map(from : a_check) + { + a_check = a; + } + if (a == a_check) + return 441; + return 0; +} + +int +test_a () +{ + int res = test_map(); + if (res != 42) + __builtin_abort (); + return res; +} + +int +test_a2 () +{ + int res = test_map_static(); + if (res != 441) + __builtin_abort (); + return res; +} diff --git a/libgomp/testsuite/libgomp.c++/target-same-name-2-b.C b/libgomp/testsuite/libgomp.c++/target-same-name-2-b.C new file mode 100644 index 00000000000..31884ba57ce --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-same-name-2-b.C @@ -0,0 +1,50 @@ +/* { dg-skip-if "" { *-*-* } } */ +/* Used by target-same-name-2.c */ + +#include + +template +int +test_map () +{ + std::complex a(2, 1), a_check; +#pragma omp target map(from : a_check) + { + a_check = a; + } + if (a == a_check) + return 42; + return 0; +} + +template +static int +test_map_static () +{ + std::complex a(-4, 5), a_check; +#pragma omp target map(from : a_check) + { + a_check = a; + } + if (a == a_check) + return 442; + return 0; +} + +int +test_b() +{ + int res = test_map(); + if (res != 42) + __builtin_abort (); + return res; +} + +int +test_b2() +{ + int res = test_map_static(); + if (res != 442) + __builtin_abort (); + return res; +} diff --git a/libgomp/testsuite/libgomp.c++/target-same-name-2.C b/libgomp/testsuite/libgomp.c++/target-same-name-2.C new file mode 100644 index 00000000000..e14d435d1ff --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-same-name-2.C @@ -0,0 +1,24 @@ +/* { dg-additional-sources "target-same-name-2-a.C target-same-name-2-b.C" } */ +/* PR middle-end/104285 */ + +/* Both files create the same symbol, which caused issues + in non-host lto1. */ + +int test_a (); +int test_a2 (); +int test_b (); +int test_b2 (); + +int +main () +{ + if (test_a () != 42) + __builtin_abort (); + if (test_a2 () != 441) + __builtin_abort (); + if (test_b () != 42) + __builtin_abort (); + if (test_b2 () != 442) + __builtin_abort (); + return 0; +} diff --git a/libgomp/testsuite/libgomp.c-c++-common/target-same-name-1-a.c b/libgomp/testsuite/libgomp.c-c++-common/target-same-name-1-a.c new file mode 100644 index 00000000000..509c238cf8d --- /dev/null +++ b/libgomp/testsuite/libgomp.c-c++-common/target-same-name-1-a.c @@ -0,0 +1,60 @@ +/* { dg-skip-if "" { *-*-* } } */ +/* Used by target-same-name-1.c */ + +static int local_link = 42; +#pragma omp declare target link(local_link) + +int decl_a_link = 123; +#pragma omp declare target link(decl_a_link) + +#pragma omp declare target +static int __attribute__ ((noinline,noclone)) +foo () +{ + return 5; +} +#pragma omp end declare target + +static int __attribute__ ((noinline,noclone)) +bar () +{ + int i; + #pragma omp target map(from:i) + i = foo (); + return i; +} + +int +one () { + return bar (); +} + +int +one_get_inc2_local_link () +{ + int res, res2; +#pragma omp target map(from: res, res2) + { + res = local_link; + local_link += 2; + res2 = local_link; + } + if (res + 2 != res2) + __builtin_abort (); + return res; +} + +int +one_get_inc3_link_a () +{ + int res, res2; +#pragma omp target map(from: res, res2) + { + res = decl_a_link; + decl_a_link += 3; + res2 = decl_a_link; + } + if (res + 3 != res2) + __builtin_abort (); + return res; +} diff --git a/libgomp/testsuite/libgomp.c-c++-common/target-same-name-1-b.c b/libgomp/testsuite/libgomp.c-c++-common/target-same-name-1-b.c new file mode 100644 index 00000000000..ce008762797 --- /dev/null +++ b/libgomp/testsuite/libgomp.c-c++-common/target-same-name-1-b.c @@ -0,0 +1,60 @@ +/* { dg-skip-if "" { *-*-* } } */ +/* Used by target-same-name-1.c */ + +static int local_link = 55; +#pragma omp declare target link(local_link) + +extern int decl_a_link; +#pragma omp declare target link(decl_a_link) + +#pragma omp declare target +static int __attribute__ ((noinline,noclone)) +foo () +{ + return 7; +} +#pragma omp end declare target + +static int __attribute__ ((noinline,noclone)) +bar () +{ + int i; + #pragma omp target map(from:i) + i = foo (); + return i; +} + +int +two () { + return bar (); +} + +int +two_get_inc4_local_link () +{ + int res, res2; +#pragma omp target map(from: res, res2) + { + res = local_link; + local_link += 4; + res2 = local_link; + } + if (res + 4 != res2) + __builtin_abort (); + return res; +} + +int +two_get_inc5_link_a () +{ + int res, res2; +#pragma omp target map(from: res, res2) + { + res = decl_a_link; + decl_a_link += 5; + res2 = decl_a_link; + } + if (res + 5 != res2) + __builtin_abort (); + return res; +} diff --git a/libgomp/testsuite/libgomp.c-c++-common/target-same-name-1.c b/libgomp/testsuite/libgomp.c-c++-common/target-same-name-1.c new file mode 100644 index 00000000000..b35d8c96ae2 --- /dev/null +++ b/libgomp/testsuite/libgomp.c-c++-common/target-same-name-1.c @@ -0,0 +1,46 @@ +/* { dg-additional-sources "target-same-name-1-a.c target-same-name-1-b.c" } */ +/* PR middle-end/104285 */ + +/* Both files create the same static symbol, which caused issues + in non-host lto1. */ + +int one (); +int two (); +int one_get_inc2_local_link (); +int two_get_inc4_local_link (); +int one_get_inc3_link_a (); +int two_get_inc5_link_a (); + +int +main () +{ + if (one () != 5) + __builtin_abort (); + if (two () != 7) + __builtin_abort (); + + if (one_get_inc2_local_link () != 42) + __builtin_abort (); + if (two_get_inc4_local_link () != 55) + __builtin_abort (); + if (one_get_inc2_local_link () != 42+2) + __builtin_abort (); + if (two_get_inc4_local_link () != 55+4) + __builtin_abort (); + + if (one_get_inc3_link_a () != 123) + __builtin_abort (); + if (two_get_inc5_link_a () != 123+3) + __builtin_abort (); + +/* FIXME: The last call did not increment the global var. */ +/* PR middle-end/105015 */ +#if 0 + if (one_get_inc3_link_a () != 123+3+5) + __builtin_abort (); + if (two_get_inc5_link_a () != 123+3+5+3) + __builtin_abort (); +#endif + + return 0; +} -- cgit v1.2.3