maxjhandsome
/
llvm-project
forked from OSchip/llvm-project
1
0
Fork 0
Code Issues Pull Requests Packages Releases Wiki Activity
llvm-project/clang/test/OpenMP/declare_target_link_codegen...

Ignoring revisions in .git-blame-ignore-revs. Click here to bypass and see the normal blame view.

90 lines
6.2 KiB
C++
Raw Normal View History

[OPENMP] Codegen for declare target with link clause. If the link clause is used on the declare target directive, the object should be linked on target or target data directives, not during the codegen. Patch adds support for this clause. llvm-svn: 328544
2018-03-27 00:40:55 +08:00
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefix HOST --check-prefix CHECK
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm-bc %s -o %t-ppc-host.bc
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck %s --check-prefix DEVICE --check-prefix CHECK
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -emit-pch -o %t
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -include-pch %t -o - | FileCheck %s --check-prefix DEVICE --check-prefix CHECK
// RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefix SIMD-ONLY
// RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm-bc %s -o %t-ppc-host.bc
// RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o -| FileCheck %s --check-prefix SIMD-ONLY
// RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -emit-pch -o %t
// RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -include-pch %t -verify -o - | FileCheck %s --check-prefix SIMD-ONLY
// expected-no-diagnostics
// SIMD-ONLY-NOT: {{__kmpc|__tgt}}
#ifndef HEADER
#define HEADER
[OPENMP] FIx processing of declare target variables. The compiler may produce unexpected error messages/crashes when declare target variables were used. Patch fixes problems with the declarations marked as declare target to or link. llvm-svn: 339805
2018-08-16 03:45:12 +08:00
// HOST-DAG: @c = external global i32,
[OpenMP] Fix declare target link implementation Summary: This patch fixes the case where variables in different compilation units or the same compilation unit are under the declare target link clause AND have the same name. This also fixes the name clash error that occurs when unified memory is activated. The changes in this patch include: - Pointers to internal variables are given unique names. - Externally visible variables are given the same name as before. - All pointer variables (external or internal) are weakly linked. Reviewers: ABataev, jdoerfert, caomhin Reviewed By: ABataev Subscribers: lebedev.ri, guansong, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D64592 llvm-svn: 367613
2019-08-02 05:15:58 +08:00
// HOST-DAG: @c_decl_tgt_ref_ptr = weak global i32* @c
// HOST-DAG: @[[D:.+]] = internal global i32 2
// HOST-DAG: @[[D_PTR:.+]] = weak global i32* @[[D]]
[OPENMP] Codegen for declare target with link clause. If the link clause is used on the declare target directive, the object should be linked on target or target data directives, not during the codegen. Patch adds support for this clause. llvm-svn: 328544
2018-03-27 00:40:55 +08:00
// DEVICE-NOT: @c =
[OpenMP] Fix declare target link implementation Summary: This patch fixes the case where variables in different compilation units or the same compilation unit are under the declare target link clause AND have the same name. This also fixes the name clash error that occurs when unified memory is activated. The changes in this patch include: - Pointers to internal variables are given unique names. - Externally visible variables are given the same name as before. - All pointer variables (external or internal) are weakly linked. Reviewers: ABataev, jdoerfert, caomhin Reviewed By: ABataev Subscribers: lebedev.ri, guansong, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D64592 llvm-svn: 367613
2019-08-02 05:15:58 +08:00
// DEVICE: @c_decl_tgt_ref_ptr = weak global i32* null
// HOST: [[SIZES:@.+]] = private unnamed_addr constant [3 x i64] [i64 4, i64 4, i64 4]
// HOST: [[MAPTYPES:@.+]] = private unnamed_addr constant [3 x i64] [i64 35, i64 531, i64 531]
[OpenMP] Add support for handling declare target to clause when unified memory is required Summary: This patch adds support for the handling of the variables under the declare target to clause. The variables in this case are handled like link variables are. A pointer is created on the host and then mapped to the device. The runtime will then copy the address of the host variable in the device pointer. Reviewers: ABataev, AlexEichenberger, caomhin Reviewed By: ABataev Subscribers: guansong, jdoerfert, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D63108 llvm-svn: 363959
2019-06-21 02:04:47 +08:00
// HOST: @.omp_offloading.entry_name{{.*}} = internal unnamed_addr constant [{{[0-9]+}} x i8] c"c_decl_tgt_ref_ptr\00"
// HOST: @.omp_offloading.entry.c_decl_tgt_ref_ptr = weak constant %struct.__tgt_offload_entry { i8* bitcast (i32** @c_decl_tgt_ref_ptr to i8*), i8* getelementptr inbounds ([{{[0-9]+}} x i8], [{{[0-9]+}} x i8]* @.omp_offloading.entry_name, i32 0, i32 0), i64 8, i32 1, i32 0 }, section ".omp_offloading.entries", align 1
[OpenMP] Fix declare target link implementation Summary: This patch fixes the case where variables in different compilation units or the same compilation unit are under the declare target link clause AND have the same name. This also fixes the name clash error that occurs when unified memory is activated. The changes in this patch include: - Pointers to internal variables are given unique names. - Externally visible variables are given the same name as before. - All pointer variables (external or internal) are weakly linked. Reviewers: ABataev, jdoerfert, caomhin Reviewed By: ABataev Subscribers: lebedev.ri, guansong, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D64592 llvm-svn: 367613
2019-08-02 05:15:58 +08:00
// DEVICE-NOT: internal unnamed_addr constant [{{[0-9]+}} x i8] c"c_{{.*}}_decl_tgt_ref_ptr\00"
// HOST: @.omp_offloading.entry_name{{.*}} = internal unnamed_addr constant [{{[0-9]+}} x i8] c"_{{.*}}d_{{.*}}_decl_tgt_ref_ptr\00"
// HOST: @.omp_offloading.entry.[[D_PTR]] = weak constant %struct.__tgt_offload_entry { i8* bitcast (i32** @[[D_PTR]] to i8*), i8* getelementptr inbounds ([{{[0-9]+}} x i8], [{{[0-9]+}} x i8]* @.omp_offloading.entry_name{{.*}}, i32 0, i32 0
[OPENMP] Codegen for declare target with link clause. If the link clause is used on the declare target directive, the object should be linked on target or target data directives, not during the codegen. Patch adds support for this clause. llvm-svn: 328544
2018-03-27 00:40:55 +08:00
extern int c;
#pragma omp declare target link(c)
[OpenMP] Fix declare target link implementation Summary: This patch fixes the case where variables in different compilation units or the same compilation unit are under the declare target link clause AND have the same name. This also fixes the name clash error that occurs when unified memory is activated. The changes in this patch include: - Pointers to internal variables are given unique names. - Externally visible variables are given the same name as before. - All pointer variables (external or internal) are weakly linked. Reviewers: ABataev, jdoerfert, caomhin Reviewed By: ABataev Subscribers: lebedev.ri, guansong, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D64592 llvm-svn: 367613
2019-08-02 05:15:58 +08:00
static int d = 2;
#pragma omp declare target link(d)
[OPENMP] Codegen for declare target with link clause. If the link clause is used on the declare target directive, the object should be linked on target or target data directives, not during the codegen. Patch adds support for this clause. llvm-svn: 328544
2018-03-27 00:40:55 +08:00
int maini1() {
int a;
#pragma omp target map(tofrom : a)
{
a = c;
[OpenMP] Fix declare target link implementation Summary: This patch fixes the case where variables in different compilation units or the same compilation unit are under the declare target link clause AND have the same name. This also fixes the name clash error that occurs when unified memory is activated. The changes in this patch include: - Pointers to internal variables are given unique names. - Externally visible variables are given the same name as before. - All pointer variables (external or internal) are weakly linked. Reviewers: ABataev, jdoerfert, caomhin Reviewed By: ABataev Subscribers: lebedev.ri, guansong, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D64592 llvm-svn: 367613
2019-08-02 05:15:58 +08:00
d++;
[OPENMP] Codegen for declare target with link clause. If the link clause is used on the declare target directive, the object should be linked on target or target data directives, not during the codegen. Patch adds support for this clause. llvm-svn: 328544
2018-03-27 00:40:55 +08:00
}
[OPENMP]Fix codegen for declare target link in target regions. If the declare target link global is used in the target region indirectly (used in the inner parallel, teams, etc. regions), we may miss this variable and it leads to incorrect codegen. llvm-svn: 355858
2019-03-12 03:51:42 +08:00
#pragma omp target
#pragma omp teams
c = a;
[OPENMP] Codegen for declare target with link clause. If the link clause is used on the declare target directive, the object should be linked on target or target data directives, not during the codegen. Patch adds support for this clause. llvm-svn: 328544
2018-03-27 00:40:55 +08:00
return 0;
}
[OpenMP] Fix declare target link implementation Summary: This patch fixes the case where variables in different compilation units or the same compilation unit are under the declare target link clause AND have the same name. This also fixes the name clash error that occurs when unified memory is activated. The changes in this patch include: - Pointers to internal variables are given unique names. - Externally visible variables are given the same name as before. - All pointer variables (external or internal) are weakly linked. Reviewers: ABataev, jdoerfert, caomhin Reviewed By: ABataev Subscribers: lebedev.ri, guansong, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D64592 llvm-svn: 367613
2019-08-02 05:15:58 +08:00
// DEVICE: define weak void @__omp_offloading_{{.*}}_{{.*}}maini1{{.*}}_l42(i32* dereferenceable{{[^,]*}}
[OpenMP] Add support for handling declare target to clause when unified memory is required Summary: This patch adds support for the handling of the variables under the declare target to clause. The variables in this case are handled like link variables are. A pointer is created on the host and then mapped to the device. The runtime will then copy the address of the host variable in the device pointer. Reviewers: ABataev, AlexEichenberger, caomhin Reviewed By: ABataev Subscribers: guansong, jdoerfert, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D63108 llvm-svn: 363959
2019-06-21 02:04:47 +08:00
// DEVICE: [[C_REF:%.+]] = load i32*, i32** @c_decl_tgt_ref_ptr,
[OPENMP] Codegen for declare target with link clause. If the link clause is used on the declare target directive, the object should be linked on target or target data directives, not during the codegen. Patch adds support for this clause. llvm-svn: 328544
2018-03-27 00:40:55 +08:00
// DEVICE: [[C:%.+]] = load i32, i32* [[C_REF]],
// DEVICE: store i32 [[C]], i32* %
// HOST: define {{.*}}i32 @{{.*}}maini1{{.*}}()
[OpenMP] Fix declare target link implementation Summary: This patch fixes the case where variables in different compilation units or the same compilation unit are under the declare target link clause AND have the same name. This also fixes the name clash error that occurs when unified memory is activated. The changes in this patch include: - Pointers to internal variables are given unique names. - Externally visible variables are given the same name as before. - All pointer variables (external or internal) are weakly linked. Reviewers: ABataev, jdoerfert, caomhin Reviewed By: ABataev Subscribers: lebedev.ri, guansong, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D64592 llvm-svn: 367613
2019-08-02 05:15:58 +08:00
// HOST: [[BASEPTRS:%.+]] = alloca [3 x i8*],
// HOST: [[PTRS:%.+]] = alloca [3 x i8*],
// HOST: getelementptr inbounds [3 x i8*], [3 x i8*]* [[BASEPTRS]], i{{[0-9]+}} 0, i{{[0-9]+}} 0
// HOST: getelementptr inbounds [3 x i8*], [3 x i8*]* [[PTRS]], i{{[0-9]+}} 0, i{{[0-9]+}} 0
// HOST: [[BP1:%.+]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[BASEPTRS]], i{{[0-9]+}} 0, i{{[0-9]+}} 1
[OPENMP] Codegen for declare target with link clause. If the link clause is used on the declare target directive, the object should be linked on target or target data directives, not during the codegen. Patch adds support for this clause. llvm-svn: 328544
2018-03-27 00:40:55 +08:00
// HOST: [[BP1_CAST:%.+]] = bitcast i8** [[BP1]] to i32***
[OpenMP] Add support for handling declare target to clause when unified memory is required Summary: This patch adds support for the handling of the variables under the declare target to clause. The variables in this case are handled like link variables are. A pointer is created on the host and then mapped to the device. The runtime will then copy the address of the host variable in the device pointer. Reviewers: ABataev, AlexEichenberger, caomhin Reviewed By: ABataev Subscribers: guansong, jdoerfert, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D63108 llvm-svn: 363959
2019-06-21 02:04:47 +08:00
// HOST: store i32** @c_decl_tgt_ref_ptr, i32*** [[BP1_CAST]],
[OpenMP] Fix declare target link implementation Summary: This patch fixes the case where variables in different compilation units or the same compilation unit are under the declare target link clause AND have the same name. This also fixes the name clash error that occurs when unified memory is activated. The changes in this patch include: - Pointers to internal variables are given unique names. - Externally visible variables are given the same name as before. - All pointer variables (external or internal) are weakly linked. Reviewers: ABataev, jdoerfert, caomhin Reviewed By: ABataev Subscribers: lebedev.ri, guansong, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D64592 llvm-svn: 367613
2019-08-02 05:15:58 +08:00
// HOST: [[P1:%.+]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[PTRS]], i{{[0-9]+}} 0, i{{[0-9]+}} 1
[OPENMP] Codegen for declare target with link clause. If the link clause is used on the declare target directive, the object should be linked on target or target data directives, not during the codegen. Patch adds support for this clause. llvm-svn: 328544
2018-03-27 00:40:55 +08:00
// HOST: [[P1_CAST:%.+]] = bitcast i8** [[P1]] to i32**
// HOST: store i32* @c, i32** [[P1_CAST]],
[OpenMP] Fix declare target link implementation Summary: This patch fixes the case where variables in different compilation units or the same compilation unit are under the declare target link clause AND have the same name. This also fixes the name clash error that occurs when unified memory is activated. The changes in this patch include: - Pointers to internal variables are given unique names. - Externally visible variables are given the same name as before. - All pointer variables (external or internal) are weakly linked. Reviewers: ABataev, jdoerfert, caomhin Reviewed By: ABataev Subscribers: lebedev.ri, guansong, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D64592 llvm-svn: 367613
2019-08-02 05:15:58 +08:00
// HOST: [[BP2:%.+]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[BASEPTRS]], i{{[0-9]+}} 0, i{{[0-9]+}} 2
// HOST: [[BP2_CAST:%.+]] = bitcast i8** [[BP2]] to i32***
// HOST: store i32** @[[D_PTR]], i32*** [[BP2_CAST]],
// HOST: [[P2:%.+]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[PTRS]], i{{[0-9]+}} 0, i{{[0-9]+}} 2
// HOST: [[P2_CAST:%.+]] = bitcast i8** [[P2]] to i32**
// HOST: store i32* @[[D]], i32** [[P2_CAST]],
// HOST: [[BP0:%.+]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[BASEPTRS]], i{{[0-9]+}} 0, i{{[0-9]+}} 0
// HOST: [[P0:%.+]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[PTRS]], i{{[0-9]+}} 0, i{{[0-9]+}} 0
// HOST: call i32 @__tgt_target(i64 -1, i8* @{{[^,]+}}, i32 3, i8** [[BP0]], i8** [[P0]], i64* getelementptr inbounds ([3 x i64], [3 x i64]* [[SIZES]], i{{[0-9]+}} 0, i{{[0-9]+}} 0), i64* getelementptr inbounds ([3 x i64], [3 x i64]* [[MAPTYPES]], i{{[0-9]+}} 0, i{{[0-9]+}} 0))
// HOST: call void @__omp_offloading_{{.*}}_{{.*}}_{{.*}}maini1{{.*}}_l42(i32* %{{[^,]+}})
// HOST: call i32 @__tgt_target_teams(i64 -1, i8* @.__omp_offloading_{{.+}}_l47.region_id, i32 2, {{.+}})
// HOST: define internal void @__omp_offloading_{{.*}}_{{.*}}maini1{{.*}}_l42(i32* dereferenceable{{.*}})
[OPENMP] Codegen for declare target with link clause. If the link clause is used on the declare target directive, the object should be linked on target or target data directives, not during the codegen. Patch adds support for this clause. llvm-svn: 328544
2018-03-27 00:40:55 +08:00
// HOST: [[C:%.*]] = load i32, i32* @c,
// HOST: store i32 [[C]], i32* %
[OpenMP] Add support for handling declare target to clause when unified memory is required Summary: This patch adds support for the handling of the variables under the declare target to clause. The variables in this case are handled like link variables are. A pointer is created on the host and then mapped to the device. The runtime will then copy the address of the host variable in the device pointer. Reviewers: ABataev, AlexEichenberger, caomhin Reviewed By: ABataev Subscribers: guansong, jdoerfert, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D63108 llvm-svn: 363959
2019-06-21 02:04:47 +08:00
// CHECK: !{i32 1, !"c_decl_tgt_ref_ptr", i32 1, i32 {{[0-9]+}}}
[OPENMP] Codegen for declare target with link clause. If the link clause is used on the declare target directive, the object should be linked on target or target data directives, not during the codegen. Patch adds support for this clause. llvm-svn: 328544
2018-03-27 00:40:55 +08:00
#endif // HEADER