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/target_parallel_codegen.cpp

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

857 lines
56 KiB
C++
Raw Normal View History

[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// Test host codegen.
[OpenMP] Update CHECK-DAG usage in target_parallel_codegen.cpp This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. There were two issues in this test: 1. There were sets of patterns for store instructions in which a pattern X could match a superset of a pattern Y. While X appeared before Y, Y's intended match appeared before X's intended match. The result was that X matched Y's intended match. Under the old overlapping behavior, Y also matched Y's intended match. Under the new non-overlapping behavior, Y had nothing left to match. This patch fixes this by gathering these sets in one place and putting the most specific patterns (Y) before the more general patterns (X). 2. The CHECK-DAG patterns involving the variables CBPADDR3 and CBPADDR4 were the same, but there was only one match in the text, so CBPADDR4 patterns had nothing to match under the new non-overlapping behavior. Moreover, a preceding related series of directives had variables (SADDR0, BPADDR0, etc.) numbered only 0 through 4, but this series had variables numbered 0 through 5. Assuming CBPADDR4's directives were not intended, this patch removes them. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54765 llvm-svn: 347351
2018-11-21 06:05:23 +08:00
// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefix CHECK --check-prefix CHECK-64
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s
[OpenMP] Update CHECK-DAG usage in target_parallel_codegen.cpp This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. There were two issues in this test: 1. There were sets of patterns for store instructions in which a pattern X could match a superset of a pattern Y. While X appeared before Y, Y's intended match appeared before X's intended match. The result was that X matched Y's intended match. Under the old overlapping behavior, Y also matched Y's intended match. Under the new non-overlapping behavior, Y had nothing left to match. This patch fixes this by gathering these sets in one place and putting the most specific patterns (Y) before the more general patterns (X). 2. The CHECK-DAG patterns involving the variables CBPADDR3 and CBPADDR4 were the same, but there was only one match in the text, so CBPADDR4 patterns had nothing to match under the new non-overlapping behavior. Moreover, a preceding related series of directives had variables (SADDR0, BPADDR0, etc.) numbered only 0 through 4, but this series had variables numbered 0 through 5. Assuming CBPADDR4's directives were not intended, this patch removes them. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54765 llvm-svn: 347351
2018-11-21 06:05:23 +08:00
// RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CHECK --check-prefix CHECK-64
// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefix CHECK --check-prefix CHECK-32
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s
[OpenMP] Update CHECK-DAG usage in target_parallel_codegen.cpp This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. There were two issues in this test: 1. There were sets of patterns for store instructions in which a pattern X could match a superset of a pattern Y. While X appeared before Y, Y's intended match appeared before X's intended match. The result was that X matched Y's intended match. Under the old overlapping behavior, Y also matched Y's intended match. Under the new non-overlapping behavior, Y had nothing left to match. This patch fixes this by gathering these sets in one place and putting the most specific patterns (Y) before the more general patterns (X). 2. The CHECK-DAG patterns involving the variables CBPADDR3 and CBPADDR4 were the same, but there was only one match in the text, so CBPADDR4 patterns had nothing to match under the new non-overlapping behavior. Moreover, a preceding related series of directives had variables (SADDR0, BPADDR0, etc.) numbered only 0 through 4, but this series had variables numbered 0 through 5. Assuming CBPADDR4's directives were not intended, this patch removes them. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54765 llvm-svn: 347351
2018-11-21 06:05:23 +08:00
// RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CHECK --check-prefix CHECK-32
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OpenMP] Update CHECK-DAG usage in target_parallel_codegen.cpp This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. There were two issues in this test: 1. There were sets of patterns for store instructions in which a pattern X could match a superset of a pattern Y. While X appeared before Y, Y's intended match appeared before X's intended match. The result was that X matched Y's intended match. Under the old overlapping behavior, Y also matched Y's intended match. Under the new non-overlapping behavior, Y had nothing left to match. This patch fixes this by gathering these sets in one place and putting the most specific patterns (Y) before the more general patterns (X). 2. The CHECK-DAG patterns involving the variables CBPADDR3 and CBPADDR4 were the same, but there was only one match in the text, so CBPADDR4 patterns had nothing to match under the new non-overlapping behavior. Moreover, a preceding related series of directives had variables (SADDR0, BPADDR0, etc.) numbered only 0 through 4, but this series had variables numbered 0 through 5. Assuming CBPADDR4's directives were not intended, this patch removes them. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54765 llvm-svn: 347351
2018-11-21 06:05:23 +08:00
// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
[OPENMP] Support for -fopenmp-simd option with compilation of simd loops only. Added support for -fopenmp-simd option that allows compilation of simd-based constructs without emission of OpenMP runtime calls. llvm-svn: 321560
2017-12-30 02:07:07 +08:00
// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s
[OpenMP] Update CHECK-DAG usage in target_parallel_codegen.cpp This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. There were two issues in this test: 1. There were sets of patterns for store instructions in which a pattern X could match a superset of a pattern Y. While X appeared before Y, Y's intended match appeared before X's intended match. The result was that X matched Y's intended match. Under the old overlapping behavior, Y also matched Y's intended match. Under the new non-overlapping behavior, Y had nothing left to match. This patch fixes this by gathering these sets in one place and putting the most specific patterns (Y) before the more general patterns (X). 2. The CHECK-DAG patterns involving the variables CBPADDR3 and CBPADDR4 were the same, but there was only one match in the text, so CBPADDR4 patterns had nothing to match under the new non-overlapping behavior. Moreover, a preceding related series of directives had variables (SADDR0, BPADDR0, etc.) numbered only 0 through 4, but this series had variables numbered 0 through 5. Assuming CBPADDR4's directives were not intended, this patch removes them. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54765 llvm-svn: 347351
2018-11-21 06:05:23 +08:00
// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
[OPENMP] Support for -fopenmp-simd option with compilation of simd loops only. Added support for -fopenmp-simd option that allows compilation of simd-based constructs without emission of OpenMP runtime calls. llvm-svn: 321560
2017-12-30 02:07:07 +08:00
// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s
[OpenMP] Update CHECK-DAG usage in target_parallel_codegen.cpp This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. There were two issues in this test: 1. There were sets of patterns for store instructions in which a pattern X could match a superset of a pattern Y. While X appeared before Y, Y's intended match appeared before X's intended match. The result was that X matched Y's intended match. Under the old overlapping behavior, Y also matched Y's intended match. Under the new non-overlapping behavior, Y had nothing left to match. This patch fixes this by gathering these sets in one place and putting the most specific patterns (Y) before the more general patterns (X). 2. The CHECK-DAG patterns involving the variables CBPADDR3 and CBPADDR4 were the same, but there was only one match in the text, so CBPADDR4 patterns had nothing to match under the new non-overlapping behavior. Moreover, a preceding related series of directives had variables (SADDR0, BPADDR0, etc.) numbered only 0 through 4, but this series had variables numbered 0 through 5. Assuming CBPADDR4's directives were not intended, this patch removes them. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54765 llvm-svn: 347351
2018-11-21 06:05:23 +08:00
// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
[OPENMP] Support for -fopenmp-simd option with compilation of simd loops only. Added support for -fopenmp-simd option that allows compilation of simd-based constructs without emission of OpenMP runtime calls. llvm-svn: 321560
2017-12-30 02:07:07 +08:00
// SIMD-ONLY0-NOT: {{__kmpc|__tgt}}
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// Test target codegen - host bc file has to be created first.
// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm-bc %s -o %t-ppc-host.bc
[OpenMP] Update CHECK-DAG usage in target_parallel_codegen.cpp This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. There were two issues in this test: 1. There were sets of patterns for store instructions in which a pattern X could match a superset of a pattern Y. While X appeared before Y, Y's intended match appeared before X's intended match. The result was that X matched Y's intended match. Under the old overlapping behavior, Y also matched Y's intended match. Under the new non-overlapping behavior, Y had nothing left to match. This patch fixes this by gathering these sets in one place and putting the most specific patterns (Y) before the more general patterns (X). 2. The CHECK-DAG patterns involving the variables CBPADDR3 and CBPADDR4 were the same, but there was only one match in the text, so CBPADDR4 patterns had nothing to match under the new non-overlapping behavior. Moreover, a preceding related series of directives had variables (SADDR0, BPADDR0, etc.) numbered only 0 through 4, but this series had variables numbered 0 through 5. Assuming CBPADDR4's directives were not intended, this patch removes them. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54765 llvm-svn: 347351
2018-11-21 06:05:23 +08:00
// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -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 TCHECK --check-prefix TCHECK-64
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o %t %s
[OpenMP] Update CHECK-DAG usage in target_parallel_codegen.cpp This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. There were two issues in this test: 1. There were sets of patterns for store instructions in which a pattern X could match a superset of a pattern Y. While X appeared before Y, Y's intended match appeared before X's intended match. The result was that X matched Y's intended match. Under the old overlapping behavior, Y also matched Y's intended match. Under the new non-overlapping behavior, Y had nothing left to match. This patch fixes this by gathering these sets in one place and putting the most specific patterns (Y) before the more general patterns (X). 2. The CHECK-DAG patterns involving the variables CBPADDR3 and CBPADDR4 were the same, but there was only one match in the text, so CBPADDR4 patterns had nothing to match under the new non-overlapping behavior. Moreover, a preceding related series of directives had variables (SADDR0, BPADDR0, etc.) numbered only 0 through 4, but this series had variables numbered 0 through 5. Assuming CBPADDR4's directives were not intended, this patch removes them. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54765 llvm-svn: 347351
2018-11-21 06:05:23 +08:00
// RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix TCHECK --check-prefix TCHECK-64
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm-bc %s -o %t-x86-host.bc
[OpenMP] Update CHECK-DAG usage in target_parallel_codegen.cpp This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. There were two issues in this test: 1. There were sets of patterns for store instructions in which a pattern X could match a superset of a pattern Y. While X appeared before Y, Y's intended match appeared before X's intended match. The result was that X matched Y's intended match. Under the old overlapping behavior, Y also matched Y's intended match. Under the new non-overlapping behavior, Y had nothing left to match. This patch fixes this by gathering these sets in one place and putting the most specific patterns (Y) before the more general patterns (X). 2. The CHECK-DAG patterns involving the variables CBPADDR3 and CBPADDR4 were the same, but there was only one match in the text, so CBPADDR4 patterns had nothing to match under the new non-overlapping behavior. Moreover, a preceding related series of directives had variables (SADDR0, BPADDR0, etc.) numbered only 0 through 4, but this series had variables numbered 0 through 5. Assuming CBPADDR4's directives were not intended, this patch removes them. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54765 llvm-svn: 347351
2018-11-21 06:05:23 +08:00
// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-x86-host.bc -o - | FileCheck %s --check-prefix TCHECK --check-prefix TCHECK-32
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -fopenmp-is-device -fopenmp-host-ir-file-path %t-x86-host.bc -o %t %s
[OpenMP] Update CHECK-DAG usage in target_parallel_codegen.cpp This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. There were two issues in this test: 1. There were sets of patterns for store instructions in which a pattern X could match a superset of a pattern Y. While X appeared before Y, Y's intended match appeared before X's intended match. The result was that X matched Y's intended match. Under the old overlapping behavior, Y also matched Y's intended match. Under the new non-overlapping behavior, Y had nothing left to match. This patch fixes this by gathering these sets in one place and putting the most specific patterns (Y) before the more general patterns (X). 2. The CHECK-DAG patterns involving the variables CBPADDR3 and CBPADDR4 were the same, but there was only one match in the text, so CBPADDR4 patterns had nothing to match under the new non-overlapping behavior. Moreover, a preceding related series of directives had variables (SADDR0, BPADDR0, etc.) numbered only 0 through 4, but this series had variables numbered 0 through 5. Assuming CBPADDR4's directives were not intended, this patch removes them. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54765 llvm-svn: 347351
2018-11-21 06:05:23 +08:00
// RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -fopenmp-is-device -fopenmp-host-ir-file-path %t-x86-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix TCHECK --check-prefix TCHECK-32
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP] Support for -fopenmp-simd option with compilation of simd loops only. Added support for -fopenmp-simd option that allows compilation of simd-based constructs without emission of OpenMP runtime calls. llvm-svn: 321560
2017-12-30 02:07:07 +08:00
// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm-bc %s -o %t-ppc-host.bc
[OpenMP] Update CHECK-DAG usage in target_parallel_codegen.cpp This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. There were two issues in this test: 1. There were sets of patterns for store instructions in which a pattern X could match a superset of a pattern Y. While X appeared before Y, Y's intended match appeared before X's intended match. The result was that X matched Y's intended match. Under the old overlapping behavior, Y also matched Y's intended match. Under the new non-overlapping behavior, Y had nothing left to match. This patch fixes this by gathering these sets in one place and putting the most specific patterns (Y) before the more general patterns (X). 2. The CHECK-DAG patterns involving the variables CBPADDR3 and CBPADDR4 were the same, but there was only one match in the text, so CBPADDR4 patterns had nothing to match under the new non-overlapping behavior. Moreover, a preceding related series of directives had variables (SADDR0, BPADDR0, etc.) numbered only 0 through 4, but this series had variables numbered 0 through 5. Assuming CBPADDR4's directives were not intended, this patch removes them. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54765 llvm-svn: 347351
2018-11-21 06:05:23 +08:00
// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -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 --check-prefix SIMD-ONLY1 %s
[OPENMP] Support for -fopenmp-simd option with compilation of simd loops only. Added support for -fopenmp-simd option that allows compilation of simd-based constructs without emission of OpenMP runtime calls. llvm-svn: 321560
2017-12-30 02:07:07 +08:00
// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o %t %s
[OpenMP] Update CHECK-DAG usage in target_parallel_codegen.cpp This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. There were two issues in this test: 1. There were sets of patterns for store instructions in which a pattern X could match a superset of a pattern Y. While X appeared before Y, Y's intended match appeared before X's intended match. The result was that X matched Y's intended match. Under the old overlapping behavior, Y also matched Y's intended match. Under the new non-overlapping behavior, Y had nothing left to match. This patch fixes this by gathering these sets in one place and putting the most specific patterns (Y) before the more general patterns (X). 2. The CHECK-DAG patterns involving the variables CBPADDR3 and CBPADDR4 were the same, but there was only one match in the text, so CBPADDR4 patterns had nothing to match under the new non-overlapping behavior. Moreover, a preceding related series of directives had variables (SADDR0, BPADDR0, etc.) numbered only 0 through 4, but this series had variables numbered 0 through 5. Assuming CBPADDR4's directives were not intended, this patch removes them. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54765 llvm-svn: 347351
2018-11-21 06:05:23 +08:00
// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY1 %s
[OPENMP] Support for -fopenmp-simd option with compilation of simd loops only. Added support for -fopenmp-simd option that allows compilation of simd-based constructs without emission of OpenMP runtime calls. llvm-svn: 321560
2017-12-30 02:07:07 +08:00
// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm-bc %s -o %t-x86-host.bc
[OpenMP] Update CHECK-DAG usage in target_parallel_codegen.cpp This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. There were two issues in this test: 1. There were sets of patterns for store instructions in which a pattern X could match a superset of a pattern Y. While X appeared before Y, Y's intended match appeared before X's intended match. The result was that X matched Y's intended match. Under the old overlapping behavior, Y also matched Y's intended match. Under the new non-overlapping behavior, Y had nothing left to match. This patch fixes this by gathering these sets in one place and putting the most specific patterns (Y) before the more general patterns (X). 2. The CHECK-DAG patterns involving the variables CBPADDR3 and CBPADDR4 were the same, but there was only one match in the text, so CBPADDR4 patterns had nothing to match under the new non-overlapping behavior. Moreover, a preceding related series of directives had variables (SADDR0, BPADDR0, etc.) numbered only 0 through 4, but this series had variables numbered 0 through 5. Assuming CBPADDR4's directives were not intended, this patch removes them. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54765 llvm-svn: 347351
2018-11-21 06:05:23 +08:00
// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-x86-host.bc -o - | FileCheck --check-prefix SIMD-ONLY1 %s
[OPENMP] Support for -fopenmp-simd option with compilation of simd loops only. Added support for -fopenmp-simd option that allows compilation of simd-based constructs without emission of OpenMP runtime calls. llvm-svn: 321560
2017-12-30 02:07:07 +08:00
// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -fopenmp-is-device -fopenmp-host-ir-file-path %t-x86-host.bc -o %t %s
[OpenMP] Update CHECK-DAG usage in target_parallel_codegen.cpp This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. There were two issues in this test: 1. There were sets of patterns for store instructions in which a pattern X could match a superset of a pattern Y. While X appeared before Y, Y's intended match appeared before X's intended match. The result was that X matched Y's intended match. Under the old overlapping behavior, Y also matched Y's intended match. Under the new non-overlapping behavior, Y had nothing left to match. This patch fixes this by gathering these sets in one place and putting the most specific patterns (Y) before the more general patterns (X). 2. The CHECK-DAG patterns involving the variables CBPADDR3 and CBPADDR4 were the same, but there was only one match in the text, so CBPADDR4 patterns had nothing to match under the new non-overlapping behavior. Moreover, a preceding related series of directives had variables (SADDR0, BPADDR0, etc.) numbered only 0 through 4, but this series had variables numbered 0 through 5. Assuming CBPADDR4's directives were not intended, this patch removes them. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54765 llvm-svn: 347351
2018-11-21 06:05:23 +08:00
// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -fopenmp-is-device -fopenmp-host-ir-file-path %t-x86-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY1 %s
[OpenMP] Ensure testing for versions 4.5 and default - Part 1 Many OpenMP Clang tests do not RUN for version 4.5 and the default version. This first patch in the series only handles test cases which do not require any modifications in the CHECK lines after adding RUN lines for default version. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D84844
2020-08-13 15:27:15 +08:00
// SIMD-ONLY1-NOT: {{__kmpc|__tgt}}
// Test host codegen.
// 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 CHECK --check-prefix CHECK-64
// RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s
// RUN: %clang_cc1 -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CHECK --check-prefix CHECK-64
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefix CHECK --check-prefix CHECK-32
// RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s
// RUN: %clang_cc1 -fopenmp -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CHECK --check-prefix CHECK-32
// RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
// RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s
// RUN: %clang_cc1 -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
// RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
// RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s
// RUN: %clang_cc1 -fopenmp-simd -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
// SIMD-ONLY0-NOT: {{__kmpc|__tgt}}
// Test target codegen - host bc file has to be created first.
// 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 TCHECK --check-prefix TCHECK-64
// RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o %t %s
// RUN: %clang_cc1 -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix TCHECK --check-prefix TCHECK-64
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm-bc %s -o %t-x86-host.bc
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-x86-host.bc -o - | FileCheck %s --check-prefix TCHECK --check-prefix TCHECK-32
// RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -fopenmp-is-device -fopenmp-host-ir-file-path %t-x86-host.bc -o %t %s
// RUN: %clang_cc1 -fopenmp -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -fopenmp-is-device -fopenmp-host-ir-file-path %t-x86-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix TCHECK --check-prefix TCHECK-32
// 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 --check-prefix SIMD-ONLY1 %s
// RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o %t %s
// RUN: %clang_cc1 -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY1 %s
// RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm-bc %s -o %t-x86-host.bc
// RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-x86-host.bc -o - | FileCheck --check-prefix SIMD-ONLY1 %s
// RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -fopenmp-is-device -fopenmp-host-ir-file-path %t-x86-host.bc -o %t %s
// RUN: %clang_cc1 -fopenmp-simd -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -fopenmp-is-device -fopenmp-host-ir-file-path %t-x86-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY1 %s
[OPENMP] Support for -fopenmp-simd option with compilation of simd loops only. Added support for -fopenmp-simd option that allows compilation of simd-based constructs without emission of OpenMP runtime calls. llvm-svn: 321560
2017-12-30 02:07:07 +08:00
// SIMD-ONLY1-NOT: {{__kmpc|__tgt}}
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// expected-no-diagnostics
#ifndef HEADER
#define HEADER
[Clang][OpenMP] Added support for nowait target in CodeGen via regular task Previously for nowait target, CG emitted a function call to `__tgt_target_nowait`, etc. However, in OpenMP RTL, these functions just directly call the no-nowait version, which means nowait is not working as expected. OpenMP specification says a target is acutally a target task, which is an untied and detachable task. It is natural to go to the direction that generates a task for a nowait target. However, OpenMP task has a problem that it must be within to a parallel region; otherwise the task will be executed immediately. As a result, if we directly wrap to a regular task, the `target nowait` outside of a parallel region is still a synchronous version. In D77609, I added the support for unshackled task in OpenMP RTL. Basically, unshackled task is a task that is not bound to any parallel region. So all nowait target will be tranformed into an unshackled task. In order to distinguish from regular task, a new flag bit is set for unshackled task. This flag will be used by RTL for later process. Since all target tasks are allocated via `__kmpc_omp_target_task_alloc`, and in current `libomptarget`, `__kmpc_omp_target_task_alloc` just calls `__kmpc_omp_task_alloc`. Therefore, we can modify the flag in `__kmpc_omp_target_task_alloc` so that we don't need to modify the FE too much. If users choose to opt out the feature, they just need to use a RTL w/o support of unshackled threads. As a result, in this patch, the `target nowait` region is simply wrapped into a regular task. Later once we have RTL support for unshackled tasks, the wrapped tasks can be executed by unshackled threads w/o changes in the FE. Reviewed By: jdoerfert Differential Revision: https://reviews.llvm.org/D78075
2020-09-26 10:10:22 +08:00
// CHECK-DAG: [[IDENT_T:%.+]] = type { i32, i32, i32, i32, i8* }
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[STR:@.+]] = private unnamed_addr constant [23 x i8] c";unknown;unknown;0;0;;\00"
[Clang][OpenMP] Added support for nowait target in CodeGen via regular task Previously for nowait target, CG emitted a function call to `__tgt_target_nowait`, etc. However, in OpenMP RTL, these functions just directly call the no-nowait version, which means nowait is not working as expected. OpenMP specification says a target is acutally a target task, which is an untied and detachable task. It is natural to go to the direction that generates a task for a nowait target. However, OpenMP task has a problem that it must be within to a parallel region; otherwise the task will be executed immediately. As a result, if we directly wrap to a regular task, the `target nowait` outside of a parallel region is still a synchronous version. In D77609, I added the support for unshackled task in OpenMP RTL. Basically, unshackled task is a task that is not bound to any parallel region. So all nowait target will be tranformed into an unshackled task. In order to distinguish from regular task, a new flag bit is set for unshackled task. This flag will be used by RTL for later process. Since all target tasks are allocated via `__kmpc_omp_target_task_alloc`, and in current `libomptarget`, `__kmpc_omp_target_task_alloc` just calls `__kmpc_omp_task_alloc`. Therefore, we can modify the flag in `__kmpc_omp_target_task_alloc` so that we don't need to modify the FE too much. If users choose to opt out the feature, they just need to use a RTL w/o support of unshackled threads. As a result, in this patch, the `target nowait` region is simply wrapped into a regular task. Later once we have RTL support for unshackled tasks, the wrapped tasks can be executed by unshackled threads w/o changes in the FE. Reviewed By: jdoerfert Differential Revision: https://reviews.llvm.org/D78075
2020-09-26 10:10:22 +08:00
// CHECK-DAG: [[ANON_T:%.+]] = type { i8 }
// CHECK-DAG: [[KMP_TASK_T_WITH_PRIVATES:%.+]] = type { [[KMP_TASK_T:%.+]] }
// CHECK-DAG: [[KMP_TASK_T]] = type { i8*, i32 (i32, i8*)*, i32, {{%[^,]+}}, {{%[^,]+}} }
[OpenMP][NFC] Reuse OMPIRBuilder `struct ident_t` handling in Clang Replace the `ident_t` handling in Clang with the methods offered by the OMPIRBuilder. This cuts down on the clang code as well as the differences between the two, making further transitions easier. Tests have changed but there should not be a real functional change. The most interesting difference is probably that we stop generating local ident_t allocations for now and just use globals. Given that this happens only with debug info, the location part of the `ident_t` is probably bigger than the test anyway. As the location part is already a global, we can avoid the allocation, memcpy, and store in favor of a constant global that is slightly bigger. This can be revisited if there are complications. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D80735
2020-05-28 22:41:01 +08:00
// CHECK-DAG: [[DEF_LOC:@.+]] = private unnamed_addr constant %struct.ident_t { i32 0, i32 2, i32 0, i32 0, i8* getelementptr inbounds ([23 x i8], [23 x i8]* [[STR]], i32 0, i32 0) }
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[TT:%.+]] = type { i64, i8 }
// CHECK-DAG: [[S1:%.+]] = type { double }
// CHECK-DAG: [[ENTTY:%.+]] = type { i8*, i8*, i[[SZ:32|64]], i32, i32 }
// TCHECK: [[ENTTY:%.+]] = type { i8*, i8*, i{{32|64}}, i32, i32 }
The patch adds support for the new map interface between clang and libomptarget. The changes in the interface are the following: device IDs are now 64-bit integers (as opposed to 32-bit) map flags are 64-bit long (used to be 32-bit) mappings for partially mapped structs are now calculated at compile time and members of partially mapped structs are flagged using the MEMBER_OF field Support for is_device_ptr on struct members was dropped - this functionality is not supported by the OpenMP standard and its implementation is technically infeasible (however, use_device_ptr on struct members works as a non-standard extension of the compiler) llvm-svn: 337468
2018-07-20 00:34:13 +08:00
// We have 8 target regions, but only 6 that actually will generate offloading
// code and have mapped arguments, and only 4 have all-constant map sizes.
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-DAG: [[SIZET2:@.+]] = private unnamed_addr constant [1 x i{{32|64}}] [i64 2]
The patch adds support for the new map interface between clang and libomptarget. The changes in the interface are the following: device IDs are now 64-bit integers (as opposed to 32-bit) map flags are 64-bit long (used to be 32-bit) mappings for partially mapped structs are now calculated at compile time and members of partially mapped structs are flagged using the MEMBER_OF field Support for is_device_ptr on struct members was dropped - this functionality is not supported by the OpenMP standard and its implementation is technically infeasible (however, use_device_ptr on struct members works as a non-standard extension of the compiler) llvm-svn: 337468
2018-07-20 00:34:13 +08:00
// CHECK-DAG: [[MAPT2:@.+]] = private unnamed_addr constant [1 x i64] [i64 800]
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-DAG: [[SIZET3:@.+]] = private unnamed_addr constant [2 x i64] [i64 4, i64 2]
The patch adds support for the new map interface between clang and libomptarget. The changes in the interface are the following: device IDs are now 64-bit integers (as opposed to 32-bit) map flags are 64-bit long (used to be 32-bit) mappings for partially mapped structs are now calculated at compile time and members of partially mapped structs are flagged using the MEMBER_OF field Support for is_device_ptr on struct members was dropped - this functionality is not supported by the OpenMP standard and its implementation is technically infeasible (however, use_device_ptr on struct members works as a non-standard extension of the compiler) llvm-svn: 337468
2018-07-20 00:34:13 +08:00
// CHECK-DAG: [[MAPT3:@.+]] = private unnamed_addr constant [2 x i64] [i64 800, i64 800]
[OPENMP]Generate correctly implicit flags for mapped data. Implicit flag must not be emitted for explicitly specified firstprivate variables, but for implicitly captured sizes of the VLAs. llvm-svn: 364575
2019-06-28 02:53:07 +08:00
// CHECK-DAG: [[MAPT4:@.+]] = private unnamed_addr constant [9 x i64] [i64 800, i64 547, i64 800, i64 547, i64 547, i64 800, i64 800, i64 547, i64 547]
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-DAG: [[SIZET5:@.+]] = private unnamed_addr constant [3 x i64] [i64 4, i64 2, i64 40]
The patch adds support for the new map interface between clang and libomptarget. The changes in the interface are the following: device IDs are now 64-bit integers (as opposed to 32-bit) map flags are 64-bit long (used to be 32-bit) mappings for partially mapped structs are now calculated at compile time and members of partially mapped structs are flagged using the MEMBER_OF field Support for is_device_ptr on struct members was dropped - this functionality is not supported by the OpenMP standard and its implementation is technically infeasible (however, use_device_ptr on struct members works as a non-standard extension of the compiler) llvm-svn: 337468
2018-07-20 00:34:13 +08:00
// CHECK-DAG: [[MAPT5:@.+]] = private unnamed_addr constant [3 x i64] [i64 800, i64 800, i64 547]
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-DAG: [[SIZET6:@.+]] = private unnamed_addr constant [4 x i64] [i64 4, i64 2, i64 1, i64 40]
The patch adds support for the new map interface between clang and libomptarget. The changes in the interface are the following: device IDs are now 64-bit integers (as opposed to 32-bit) map flags are 64-bit long (used to be 32-bit) mappings for partially mapped structs are now calculated at compile time and members of partially mapped structs are flagged using the MEMBER_OF field Support for is_device_ptr on struct members was dropped - this functionality is not supported by the OpenMP standard and its implementation is technically infeasible (however, use_device_ptr on struct members works as a non-standard extension of the compiler) llvm-svn: 337468
2018-07-20 00:34:13 +08:00
// CHECK-DAG: [[MAPT6:@.+]] = private unnamed_addr constant [4 x i64] [i64 800, i64 800, i64 800, i64 547]
[OPENMP]Fix PR48076: mapping of data member pointer. If the data member pointer is mapped, the compiler tries to optimize the mapping of such data by discarding explicit mapping flags and trying to emit combined data instead. In some cases, this optimization is not quite correctly implemented and it leads to a program crash at the runtime. Instead, if the data member is mapped, just emit it as is and do not emit combined mapping flags for it. Differential Revision: https://reviews.llvm.org/D91552
2020-11-17 01:43:53 +08:00
// CHECK-DAG: [[MAPT7:@.+]] = private unnamed_addr constant [5 x i64] [i64 547, i64 800, i64 800, i64 800, i64 547]
[OPENMP, NVPTX] Fix linkage of the global entries. The linkage of the global entries must be weak to enable support of redefinition of the same target regions in multiple compilation units. llvm-svn: 331768
2018-05-08 22:16:57 +08:00
// CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0
// CHECK-DAG: @{{.*}} = weak constant i8 0
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP] Added emission of offloading data sections for declare target variables. Added emission of the offloading data sections for the variables within declare target regions + fixes emission of the declare target variables marked as declare target not within the declare target region. llvm-svn: 328888
2018-03-31 02:31:07 +08:00
// TCHECK: @{{.+}} = weak constant [[ENTTY]]
// TCHECK: @{{.+}} = weak constant [[ENTTY]]
// TCHECK: @{{.+}} = weak constant [[ENTTY]]
// TCHECK: @{{.+}} = weak constant [[ENTTY]]
// TCHECK: @{{.+}} = weak constant [[ENTTY]]
// TCHECK: @{{.+}} = weak constant [[ENTTY]]
// TCHECK: @{{.+}} = weak constant [[ENTTY]]
// TCHECK-NOT: @{{.+}} = weak constant [[ENTTY]]
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// Check target registration is registered as a Ctor.
[Clang][OpenMP Offload] Move offload registration code to the wrapper The final list of OpenMP offload targets becomes known only at the link time and since offload registration code depends on the targets list it makes sense to delay offload registration code generation to the link time instead of adding it to the host part of every fat object. This patch moves offload registration code generation from clang to the offload wrapper tool. This is the last part of the OpenMP linker script elimination patch https://reviews.llvm.org/D64943 Differential Revision: https://reviews.llvm.org/D68746 llvm-svn: 374937
2019-10-16 02:42:47 +08:00
// CHECK: appending global [1 x { i32, void ()*, i8* }] [{ i32, void ()*, i8* } { i32 0, void ()* @.omp_offloading.requires_reg, i8* null }]
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
template<typename tx, typename ty>
struct TT{
tx X;
ty Y;
};
// CHECK: define {{.*}}[[FOO:@.+]](
int foo(int n) {
int a = 0;
short aa = 0;
float b[10];
float bn[n];
double c[5][10];
double cn[5][n];
TT<long long, char> d;
[Clang][OpenMP] Added support for nowait target in CodeGen via regular task Previously for nowait target, CG emitted a function call to `__tgt_target_nowait`, etc. However, in OpenMP RTL, these functions just directly call the no-nowait version, which means nowait is not working as expected. OpenMP specification says a target is acutally a target task, which is an untied and detachable task. It is natural to go to the direction that generates a task for a nowait target. However, OpenMP task has a problem that it must be within to a parallel region; otherwise the task will be executed immediately. As a result, if we directly wrap to a regular task, the `target nowait` outside of a parallel region is still a synchronous version. In D77609, I added the support for unshackled task in OpenMP RTL. Basically, unshackled task is a task that is not bound to any parallel region. So all nowait target will be tranformed into an unshackled task. In order to distinguish from regular task, a new flag bit is set for unshackled task. This flag will be used by RTL for later process. Since all target tasks are allocated via `__kmpc_omp_target_task_alloc`, and in current `libomptarget`, `__kmpc_omp_target_task_alloc` just calls `__kmpc_omp_task_alloc`. Therefore, we can modify the flag in `__kmpc_omp_target_task_alloc` so that we don't need to modify the FE too much. If users choose to opt out the feature, they just need to use a RTL w/o support of unshackled threads. As a result, in this patch, the `target nowait` region is simply wrapped into a regular task. Later once we have RTL support for unshackled tasks, the wrapped tasks can be executed by unshackled threads w/o changes in the FE. Reviewed By: jdoerfert Differential Revision: https://reviews.llvm.org/D78075
2020-09-26 10:10:22 +08:00
// CHECK-32: [[TASK:%.+]] = call i8* @__kmpc_omp_target_task_alloc([[IDENT_T]]* {{[^,]+}}, i32 %0, i32 1, i32 20, i32 1, i32 (i32, i8*)* bitcast (i32 (i32, [[KMP_TASK_T_WITH_PRIVATES:%.+]]*)* [[OMP_TASK_ENTRY:@.+]] to i32 (i32, i8*)*), i64 -1)
// CHECK-64: [[TASK:%.+]] = call i8* @__kmpc_omp_target_task_alloc([[IDENT_T]]* {{[^,]+}}, i32 %0, i32 1, i64 40, i64 1, i32 (i32, i8*)* bitcast (i32 (i32, [[KMP_TASK_T_WITH_PRIVATES:%.+]]*)* [[OMP_TASK_ENTRY:@.+]] to i32 (i32, i8*)*), i64 -1)
// CHECK: call i32 @__kmpc_omp_task([[IDENT_T]]* {{[^,]+}}, i32 {{%[^,]+}}, i8* [[TASK]])
[OPENMP] Add codegen for `nowait` clause in target directives. Added basic codegen for `nowait` clauses in target-based directives. llvm-svn: 320613
2017-12-14 05:04:20 +08:00
#pragma omp target parallel nowait
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
{
}
// CHECK: call void [[HVT1:@.+]](i[[SZ]] {{[^,]+}})
#pragma omp target parallel if(target: 0)
{
a += 1;
}
[OpenMP] Add Location Fields to Libomptarget Runtime for Debugging Summary: Add support for passing source locations to libomptarget runtime functions using the ident_t struct present in the rest of the libomp API. This will allow the runtime system to give much more insightful error messages and debugging values. Reviewers: jdoerfert grokos Differential Revision: https://reviews.llvm.org/D87946
2020-11-20 00:56:59 +08:00
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(%struct.ident_t* @{{.+}}, i64 -1, i8* @{{[^,]+}}, i32 1, i8** [[BP:%[^,]+]], i8** [[P:%[^,]+]], i64* getelementptr inbounds ([1 x i64], [1 x i64]* [[SIZET2]], i32 0, i32 0), i64* getelementptr inbounds ([1 x i64], [1 x i64]* [[MAPT2]], i32 0, i32 0), i8** null, i8** null, i32 1, i32 0)
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BP]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[BPR:%[^,]+]], i32 0, i32 0
// CHECK-DAG: [[P]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[PR:%[^,]+]], i32 0, i32 0
// CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[BPR]], i32 0, i32 [[IDX0:[0-9]+]]
// CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[PR]], i32 0, i32 [[IDX0]]
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: [[CBPADDR0:%.+]] = bitcast i8** [[BPADDR0]] to i[[SZ]]*
// CHECK-DAG: [[CPADDR0:%.+]] = bitcast i8** [[PADDR0]] to i[[SZ]]*
// CHECK-DAG: store i[[SZ]] [[VAL0:%.+]], i[[SZ]]* [[CBPADDR0]],
// CHECK-DAG: store i[[SZ]] [[VAL0]], i[[SZ]]* [[CPADDR0]],
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP] Simplify codegen for non-offloading code. Simplified and generalized codegen for non-offloading part that works if offloading is failed or condition of the `if` clause is `false`. llvm-svn: 314670
2017-10-02 22:20:58 +08:00
// CHECK-NEXT: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:[^,]+]], label %[[END:[^,]+]]
// CHECK: [[FAIL]]
// CHECK: call void [[HVT2:@.+]](i[[SZ]] {{[^,]+}})
// CHECK-NEXT: br label %[[END]]
// CHECK: [[END]]
#pragma omp target parallel if(target: 1)
{
aa += 1;
[OPENMP]Allow cancellation constructs in target parallel regions. Summary: omp cancellation point parallel and omp cancel parallel directives are allowed in target paralle regions. Reviewers: jdoerfert Subscribers: yaxunl, guansong, caomhin, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D78941
2020-04-27 23:37:35 +08:00
#pragma omp cancel parallel
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
}
// CHECK: [[IF:%.+]] = icmp sgt i32 {{[^,]+}}, 10
// CHECK: br i1 [[IF]], label %[[IFTHEN:[^,]+]], label %[[IFELSE:[^,]+]]
// CHECK: [[IFTHEN]]
[OpenMP] Add Location Fields to Libomptarget Runtime for Debugging Summary: Add support for passing source locations to libomptarget runtime functions using the ident_t struct present in the rest of the libomp API. This will allow the runtime system to give much more insightful error messages and debugging values. Reviewers: jdoerfert grokos Differential Revision: https://reviews.llvm.org/D87946
2020-11-20 00:56:59 +08:00
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(%struct.ident_t* @{{.+}}, i64 -1, i8* @{{[^,]+}}, i32 2, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* getelementptr inbounds ([2 x i64], [2 x i64]* [[SIZET3]], i32 0, i32 0), i64* getelementptr inbounds ([2 x i64], [2 x i64]* [[MAPT3]], i32 0, i32 0), i8** null, i8** null, i32 1, i32 0)
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPR]] = getelementptr inbounds [2 x i8*], [2 x i8*]* [[BP:%[^,]+]], i32 0, i32 0
// CHECK-DAG: [[PR]] = getelementptr inbounds [2 x i8*], [2 x i8*]* [[P:%[^,]+]], i32 0, i32 0
// CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [2 x i8*], [2 x i8*]* [[BP]], i32 0, i32 0
// CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [2 x i8*], [2 x i8*]* [[P]], i32 0, i32 0
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: [[CBPADDR0:%.+]] = bitcast i8** [[BPADDR0]] to i[[SZ]]*
// CHECK-DAG: [[CPADDR0:%.+]] = bitcast i8** [[PADDR0]] to i[[SZ]]*
// CHECK-DAG: store i[[SZ]] [[VAL0:%.+]], i[[SZ]]* [[CBPADDR0]],
// CHECK-DAG: store i[[SZ]] [[VAL0]], i[[SZ]]* [[CPADDR0]],
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [2 x i8*], [2 x i8*]* [[BP]], i32 0, i32 1
// CHECK-DAG: [[PADDR1:%.+]] = getelementptr inbounds [2 x i8*], [2 x i8*]* [[P]], i32 0, i32 1
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: [[CBPADDR1:%.+]] = bitcast i8** [[BPADDR1]] to i[[SZ]]*
// CHECK-DAG: [[CPADDR1:%.+]] = bitcast i8** [[PADDR1]] to i[[SZ]]*
// CHECK-DAG: store i[[SZ]] [[VAL1:%.+]], i[[SZ]]* [[CBPADDR1]],
// CHECK-DAG: store i[[SZ]] [[VAL1]], i[[SZ]]* [[CPADDR1]],
[OPENMP] Simplify codegen for non-offloading code. Simplified and generalized codegen for non-offloading part that works if offloading is failed or condition of the `if` clause is `false`. llvm-svn: 314670
2017-10-02 22:20:58 +08:00
// CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
// CHECK: [[FAIL]]
// CHECK: call void [[HVT3:@.+]]({{[^,]+}}, {{[^,]+}})
// CHECK-NEXT: br label %[[END]]
// CHECK: [[END]]
[OPENMP] Simplify codegen for non-offloading code. Simplified and generalized codegen for non-offloading part that works if offloading is failed or condition of the `if` clause is `false`. llvm-svn: 314670
2017-10-02 22:20:58 +08:00
// CHECK-NEXT: br label %[[IFEND:.+]]
// CHECK: [[IFELSE]]
// CHECK: call void [[HVT3]]({{[^,]+}}, {{[^,]+}})
// CHECK-NEXT: br label %[[IFEND]]
// CHECK: [[IFEND]]
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
#pragma omp target parallel if(target: n>10)
{
a += 1;
aa += 1;
}
// We capture 3 VLA sizes in this target region
// CHECK-64: [[A_VAL:%.+]] = load i32, i32* %{{.+}},
// CHECK-64: [[A_ADDR:%.+]] = bitcast i[[SZ]]* [[A_CADDR:%.+]] to i32*
// CHECK-64: store i32 [[A_VAL]], i32* [[A_ADDR]],
// CHECK-64: [[A_CVAL:%.+]] = load i[[SZ]], i[[SZ]]* [[A_CADDR]],
// CHECK-32: [[A_VAL:%.+]] = load i32, i32* %{{.+}},
// CHECK-32: store i32 [[A_VAL]], i32* [[A_CADDR:%.+]],
// CHECK-32: [[A_CVAL:%.+]] = load i[[SZ]], i[[SZ]]* [[A_CADDR]],
[OPENMP] Add codegen for `depend` clauses on `target` directive. Added basic support for codegen of `depend` clauses on `target` directive. llvm-svn: 322501
2018-01-16 03:06:12 +08:00
// CHECK: [[IF:%.+]] = icmp sgt i32 {{[^,]+}}, 20
// CHECK: br i1 [[IF]], label %[[TRY:[^,]+]], label %[[FAIL:[^,]+]]
// CHECK: [[TRY]]
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-64: [[BNSIZE:%.+]] = mul nuw i64 [[VLA0:%.+]], 4
// CHECK-32: [[BNSZSIZE:%.+]] = mul nuw i32 [[VLA0:%.+]], 4
// CHECK-32: [[BNSIZE:%.+]] = sext i32 [[BNSZSIZE]] to i64
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK: [[CNELEMSIZE2:%.+]] = mul nuw i[[SZ]] 5, [[VLA1:%.+]]
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-64: [[CNSIZE:%.+]] = mul nuw i64 [[CNELEMSIZE2]], 8
// CHECK-32: [[CNSZSIZE:%.+]] = mul nuw i32 [[CNELEMSIZE2]], 8
// CHECK-32: [[CNSIZE:%.+]] = sext i32 [[CNSZSIZE]] to i64
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OpenMP] Add Location Fields to Libomptarget Runtime for Debugging Summary: Add support for passing source locations to libomptarget runtime functions using the ident_t struct present in the rest of the libomp API. This will allow the runtime system to give much more insightful error messages and debugging values. Reviewers: jdoerfert grokos Differential Revision: https://reviews.llvm.org/D87946
2020-11-20 00:56:59 +08:00
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(%struct.ident_t* @{{.+}}, i64 -1, i8* @{{[^,]+}}, i32 9, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([9 x i64], [9 x i64]* [[MAPT4]], i32 0, i32 0), i8** null, i8** null, i32 1, i32 0)
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPR]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[BP:%[^,]+]], i32 0, i32 0
// CHECK-DAG: [[PR]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[P:%[^,]+]], i32 0, i32 0
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-DAG: [[SR]] = getelementptr inbounds [9 x i64], [9 x i64]* [[S:%[^,]+]], i32 0, i32 0
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-DAG: [[SADDR0:%.+]] = getelementptr inbounds [9 x i64], [9 x i64]* [[S]], i32 0, i32 [[IDX0:[0-9]+]]
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[BP]], i32 0, i32 [[IDX0]]
// CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[P]], i32 0, i32 [[IDX0]]
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-DAG: [[SADDR1:%.+]] = getelementptr inbounds [9 x i64], [9 x i64]* [[S]], i32 0, i32 [[IDX1:[0-9]+]]
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[BP]], i32 0, i32 [[IDX1]]
// CHECK-DAG: [[PADDR1:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[P]], i32 0, i32 [[IDX1]]
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-DAG: [[SADDR2:%.+]] = getelementptr inbounds [9 x i64], [9 x i64]* [[S]], i32 0, i32 [[IDX2:[0-9]+]]
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[BP]], i32 0, i32 [[IDX2]]
// CHECK-DAG: [[PADDR2:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[P]], i32 0, i32 [[IDX2]]
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-DAG: [[SADDR3:%.+]] = getelementptr inbounds [9 x i64], [9 x i64]* [[S]], i32 0, i32 [[IDX3:[0-9]+]]
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[BP]], i32 0, i32 [[IDX3]]
// CHECK-DAG: [[PADDR3:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[P]], i32 0, i32 [[IDX3]]
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-DAG: [[SADDR4:%.+]] = getelementptr inbounds [9 x i64], [9 x i64]* [[S]], i32 0, i32 [[IDX4:[0-9]+]]
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[BP]], i32 0, i32 [[IDX4]]
// CHECK-DAG: [[PADDR4:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[P]], i32 0, i32 [[IDX4]]
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-DAG: [[SADDR5:%.+]] = getelementptr inbounds [9 x i64], [9 x i64]* [[S]], i32 0, i32 [[IDX5:[0-9]+]]
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPADDR5:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[BP]], i32 0, i32 [[IDX5]]
// CHECK-DAG: [[PADDR5:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[P]], i32 0, i32 [[IDX5]]
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-DAG: [[SADDR6:%.+]] = getelementptr inbounds [9 x i64], [9 x i64]* [[S]], i32 0, i32 [[IDX6:[0-9]+]]
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPADDR6:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[BP]], i32 0, i32 [[IDX6]]
// CHECK-DAG: [[PADDR6:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[P]], i32 0, i32 [[IDX6]]
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-DAG: [[SADDR7:%.+]] = getelementptr inbounds [9 x i64], [9 x i64]* [[S]], i32 0, i32 [[IDX7:[0-9]+]]
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPADDR7:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[BP]], i32 0, i32 [[IDX7]]
// CHECK-DAG: [[PADDR7:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[P]], i32 0, i32 [[IDX7]]
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-DAG: [[SADDR8:%.+]] = getelementptr inbounds [9 x i64], [9 x i64]* [[S]], i32 0, i32 [[IDX8:[0-9]+]]
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPADDR8:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[BP]], i32 0, i32 [[IDX8]]
// CHECK-DAG: [[PADDR8:%.+]] = getelementptr inbounds [9 x i8*], [9 x i8*]* [[P]], i32 0, i32 [[IDX8]]
// The names below are not necessarily consistent with the names used for the
// addresses above as some are repeated.
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CBPADDR0:%.+]],
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CPADDR0:%.+]],
// CHECK-DAG: [[CBPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: store i[[SZ]] [[VLA1]], i[[SZ]]* [[CBPADDR1:%.+]],
// CHECK-DAG: store i[[SZ]] [[VLA1]], i[[SZ]]* [[CPADDR1:%.+]],
// CHECK-DAG: [[CBPADDR1]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR1]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: store i[[SZ]] 5, i[[SZ]]* [[CBPADDR2:%.+]],
// CHECK-DAG: store i[[SZ]] 5, i[[SZ]]* [[CPADDR2:%.+]],
// CHECK-DAG: [[CBPADDR2]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR2]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: store i[[SZ]] [[A_CVAL]], i[[SZ]]* [[CBPADDR3:%.+]],
// CHECK-DAG: store i[[SZ]] [[A_CVAL]], i[[SZ]]* [[CPADDR3:%.+]],
// CHECK-DAG: [[CBPADDR3]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR3]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: store [10 x float]* %{{.+}}, [10 x float]** [[CBPADDR4:%.+]],
// CHECK-DAG: store [10 x float]* %{{.+}}, [10 x float]** [[CPADDR4:%.+]],
// CHECK-DAG: [[CBPADDR4]] = bitcast i8** {{%[^,]+}} to [10 x float]**
// CHECK-DAG: [[CPADDR4]] = bitcast i8** {{%[^,]+}} to [10 x float]**
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: store float* %{{.+}}, float** [[CBPADDR5:%.+]],
// CHECK-DAG: store float* %{{.+}}, float** [[CPADDR5:%.+]],
// CHECK-DAG: [[CBPADDR5]] = bitcast i8** {{%[^,]+}} to float**
// CHECK-DAG: [[CPADDR5]] = bitcast i8** {{%[^,]+}} to float**
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: store [5 x [10 x double]]* %{{.+}}, [5 x [10 x double]]** [[CBPADDR6:%.+]],
// CHECK-DAG: store [5 x [10 x double]]* %{{.+}}, [5 x [10 x double]]** [[CPADDR6:%.+]],
// CHECK-DAG: [[CBPADDR6]] = bitcast i8** {{%[^,]+}} to [5 x [10 x double]]**
// CHECK-DAG: [[CPADDR6]] = bitcast i8** {{%[^,]+}} to [5 x [10 x double]]**
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: store double* %{{.+}}, double** [[CBPADDR7:%.+]],
// CHECK-DAG: store double* %{{.+}}, double** [[CPADDR7:%.+]],
// CHECK-DAG: [[CBPADDR7]] = bitcast i8** {{%[^,]+}} to double**
// CHECK-DAG: [[CPADDR7]] = bitcast i8** {{%[^,]+}} to double**
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: store [[TT]]* %{{.+}}, [[TT]]** [[CBPADDR8:%.+]],
// CHECK-DAG: store [[TT]]* %{{.+}}, [[TT]]** [[CPADDR8:%.+]],
// CHECK-DAG: [[CBPADDR8]] = bitcast i8** {{%[^,]+}} to [[TT]]**
// CHECK-DAG: [[CPADDR8]] = bitcast i8** {{%[^,]+}} to [[TT]]**
[OpenMP] Update CHECK-DAG usage in target_parallel_codegen.cpp This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. There were two issues in this test: 1. There were sets of patterns for store instructions in which a pattern X could match a superset of a pattern Y. While X appeared before Y, Y's intended match appeared before X's intended match. The result was that X matched Y's intended match. Under the old overlapping behavior, Y also matched Y's intended match. Under the new non-overlapping behavior, Y had nothing left to match. This patch fixes this by gathering these sets in one place and putting the most specific patterns (Y) before the more general patterns (X). 2. The CHECK-DAG patterns involving the variables CBPADDR3 and CBPADDR4 were the same, but there was only one match in the text, so CBPADDR4 patterns had nothing to match under the new non-overlapping behavior. Moreover, a preceding related series of directives had variables (SADDR0, BPADDR0, etc.) numbered only 0 through 4, but this series had variables numbered 0 through 5. Assuming CBPADDR4's directives were not intended, this patch removes them. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54765 llvm-svn: 347351
2018-11-21 06:05:23 +08:00
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-DAG: store i64 4, i64* {{%[^,]+}}
// CHECK-DAG: store i64 [[CNSIZE]], i64* {{%[^,]+}}
// CHECK-DAG: store i64 [[BNSIZE]], i64* {{%[^,]+}}
// CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}}
// CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}}
// CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}}
// CHECK-DAG: store i64 40, i64* {{%[^,]+}}
// CHECK-DAG: store i64 400, i64* {{%[^,]+}}
// CHECK-DAG: store i64 {{12|16}}, i64* {{%[^,]+}}
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP] Simplify codegen for non-offloading code. Simplified and generalized codegen for non-offloading part that works if offloading is failed or condition of the `if` clause is `false`. llvm-svn: 314670
2017-10-02 22:20:58 +08:00
// CHECK-NEXT: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:[^,]+]], label %[[END:[^,]+]]
// CHECK: [[FAIL]]
// CHECK: call void [[HVT4:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}})
// CHECK-NEXT: br label %[[END]]
// CHECK: [[END]]
#pragma omp target parallel if(target: n>20)
{
a += 1;
b[2] += 1.0;
bn[3] += 1.0;
c[1][2] += 1.0;
cn[1][3] += 1.0;
d.X += 1;
d.Y += 1;
}
return a;
}
// Check that the offloading functions are emitted and that the arguments are
// correct and loaded correctly for the target regions in foo().
[Clang][OpenMP] Added support for nowait target in CodeGen via regular task Previously for nowait target, CG emitted a function call to `__tgt_target_nowait`, etc. However, in OpenMP RTL, these functions just directly call the no-nowait version, which means nowait is not working as expected. OpenMP specification says a target is acutally a target task, which is an untied and detachable task. It is natural to go to the direction that generates a task for a nowait target. However, OpenMP task has a problem that it must be within to a parallel region; otherwise the task will be executed immediately. As a result, if we directly wrap to a regular task, the `target nowait` outside of a parallel region is still a synchronous version. In D77609, I added the support for unshackled task in OpenMP RTL. Basically, unshackled task is a task that is not bound to any parallel region. So all nowait target will be tranformed into an unshackled task. In order to distinguish from regular task, a new flag bit is set for unshackled task. This flag will be used by RTL for later process. Since all target tasks are allocated via `__kmpc_omp_target_task_alloc`, and in current `libomptarget`, `__kmpc_omp_target_task_alloc` just calls `__kmpc_omp_task_alloc`. Therefore, we can modify the flag in `__kmpc_omp_target_task_alloc` so that we don't need to modify the FE too much. If users choose to opt out the feature, they just need to use a RTL w/o support of unshackled threads. As a result, in this patch, the `target nowait` region is simply wrapped into a regular task. Later once we have RTL support for unshackled tasks, the wrapped tasks can be executed by unshackled threads w/o changes in the FE. Reviewed By: jdoerfert Differential Revision: https://reviews.llvm.org/D78075
2020-09-26 10:10:22 +08:00
// CHECK: define internal void [[HVT0:@.+]]()
[OPENMP] General code improvements. llvm-svn: 330140
2018-04-17 01:59:34 +08:00
// CHECK: call {{.*}}void (%struct.ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%struct.ident_t* [[DEF_LOC]], i32 0, void (i32*, i32*, ...)* bitcast (void (i32*, i32*)* [[OMP_OUTLINED:@.+]] to void (i32*, i32*, ...)*))
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
//
//
// CHECK: define internal {{.*}}void [[OMP_OUTLINED]](i32* noalias %.global_tid., i32* noalias %.bound_tid.)
// CHECK: ret void
// CHECK-NEXT: }
[Clang][OpenMP] Added support for nowait target in CodeGen via regular task Previously for nowait target, CG emitted a function call to `__tgt_target_nowait`, etc. However, in OpenMP RTL, these functions just directly call the no-nowait version, which means nowait is not working as expected. OpenMP specification says a target is acutally a target task, which is an untied and detachable task. It is natural to go to the direction that generates a task for a nowait target. However, OpenMP task has a problem that it must be within to a parallel region; otherwise the task will be executed immediately. As a result, if we directly wrap to a regular task, the `target nowait` outside of a parallel region is still a synchronous version. In D77609, I added the support for unshackled task in OpenMP RTL. Basically, unshackled task is a task that is not bound to any parallel region. So all nowait target will be tranformed into an unshackled task. In order to distinguish from regular task, a new flag bit is set for unshackled task. This flag will be used by RTL for later process. Since all target tasks are allocated via `__kmpc_omp_target_task_alloc`, and in current `libomptarget`, `__kmpc_omp_target_task_alloc` just calls `__kmpc_omp_task_alloc`. Therefore, we can modify the flag in `__kmpc_omp_target_task_alloc` so that we don't need to modify the FE too much. If users choose to opt out the feature, they just need to use a RTL w/o support of unshackled threads. As a result, in this patch, the `target nowait` region is simply wrapped into a regular task. Later once we have RTL support for unshackled tasks, the wrapped tasks can be executed by unshackled threads w/o changes in the FE. Reviewed By: jdoerfert Differential Revision: https://reviews.llvm.org/D78075
2020-09-26 10:10:22 +08:00
// CHECK: define internal {{.*}}i32 [[OMP_TASK_ENTRY]](i32 {{.*}}%0, [[KMP_TASK_T_WITH_PRIVATES]]* noalias %1)
[OpenMP] Add Location Fields to Libomptarget Runtime for Debugging Summary: Add support for passing source locations to libomptarget runtime functions using the ident_t struct present in the rest of the libomp API. This will allow the runtime system to give much more insightful error messages and debugging values. Reviewers: jdoerfert grokos Differential Revision: https://reviews.llvm.org/D87946
2020-11-20 00:56:59 +08:00
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_nowait_mapper(%struct.ident_t* @{{.+}}, i64 -1, i8* {{@[^,]+}}, i32 0, i8** null, i8** null, i64* null, i64* null, i8** null, i8** null, i32 1, i32 0)
[Clang][OpenMP] Added support for nowait target in CodeGen via regular task Previously for nowait target, CG emitted a function call to `__tgt_target_nowait`, etc. However, in OpenMP RTL, these functions just directly call the no-nowait version, which means nowait is not working as expected. OpenMP specification says a target is acutally a target task, which is an untied and detachable task. It is natural to go to the direction that generates a task for a nowait target. However, OpenMP task has a problem that it must be within to a parallel region; otherwise the task will be executed immediately. As a result, if we directly wrap to a regular task, the `target nowait` outside of a parallel region is still a synchronous version. In D77609, I added the support for unshackled task in OpenMP RTL. Basically, unshackled task is a task that is not bound to any parallel region. So all nowait target will be tranformed into an unshackled task. In order to distinguish from regular task, a new flag bit is set for unshackled task. This flag will be used by RTL for later process. Since all target tasks are allocated via `__kmpc_omp_target_task_alloc`, and in current `libomptarget`, `__kmpc_omp_target_task_alloc` just calls `__kmpc_omp_task_alloc`. Therefore, we can modify the flag in `__kmpc_omp_target_task_alloc` so that we don't need to modify the FE too much. If users choose to opt out the feature, they just need to use a RTL w/o support of unshackled threads. As a result, in this patch, the `target nowait` region is simply wrapped into a regular task. Later once we have RTL support for unshackled tasks, the wrapped tasks can be executed by unshackled threads w/o changes in the FE. Reviewed By: jdoerfert Differential Revision: https://reviews.llvm.org/D78075
2020-09-26 10:10:22 +08:00
// CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
// CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:[^,]+]], label %[[END:[^,]+]]
// CHECK: [[FAIL]]
// CHECK: call void [[HVT0]]()
// CHECK-NEXT: br label %[[END]]
// CHECK: [[END]]
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK: define internal void [[HVT1]](i[[SZ]] %{{.+}})
// Create stack storage and store argument in there.
// CHECK: [[AA_ADDR:%.+]] = alloca i[[SZ]], align
// CHECK: [[AA_CASTED:%.+]] = alloca i[[SZ]], align
// CHECK: store i[[SZ]] %{{.+}}, i[[SZ]]* [[AA_ADDR]], align
// CHECK-64: [[AA_CADDR:%.+]] = bitcast i[[SZ]]* [[AA_ADDR]] to i32*
// CHECK-64: [[AA:%.+]] = load i32, i32* [[AA_CADDR]], align
// CHECK-32: [[AA:%.+]] = load i32, i32* [[AA_ADDR]], align
// CHECK-64: [[AA_C:%.+]] = bitcast i[[SZ]]* [[AA_CASTED]] to i32*
// CHECK-64: store i32 [[AA]], i32* [[AA_C]], align
// CHECK-32: store i32 [[AA]], i32* [[AA_CASTED]], align
// CHECK: [[PARAM:%.+]] = load i[[SZ]], i[[SZ]]* [[AA_CASTED]], align
[OPENMP] General code improvements. llvm-svn: 330140
2018-04-17 01:59:34 +08:00
// CHECK: call {{.*}}void (%struct.ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%struct.ident_t* [[DEF_LOC]], i32 1, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, i[[SZ]])* [[OMP_OUTLINED1:@.+]] to void (i32*, i32*, ...)*), i[[SZ]] [[PARAM]])
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
//
//
// CHECK: define internal {{.*}}void [[OMP_OUTLINED1]](i32* noalias %.global_tid., i32* noalias %.bound_tid., i[[SZ]] %{{.+}})
// CHECK: [[AA_ADDR:%.+]] = alloca i[[SZ]], align
// CHECK: store i[[SZ]] %{{.+}}, i[[SZ]]* [[AA_ADDR]], align
// CHECK-64: [[AA_CADDR:%.+]] = bitcast i[[SZ]]* [[AA_ADDR]] to i32*
// CHECK-64: [[AA:%.+]] = load i32, i32* [[AA_CADDR]], align
// CHECK-32: [[AA:%.+]] = load i32, i32* [[AA_ADDR]], align
// CHECK: ret void
// CHECK-NEXT: }
// CHECK: define internal void [[HVT2]](i[[SZ]] %{{.+}})
// Create stack storage and store argument in there.
// CHECK: [[AA_ADDR:%.+]] = alloca i[[SZ]], align
// CHECK: [[AA_CASTED:%.+]] = alloca i[[SZ]], align
// CHECK: store i[[SZ]] %{{.+}}, i[[SZ]]* [[AA_ADDR]], align
// CHECK: [[AA_CADDR:%.+]] = bitcast i[[SZ]]* [[AA_ADDR]] to i16*
// CHECK: [[AA:%.+]] = load i16, i16* [[AA_CADDR]], align
// CHECK: [[AA_C:%.+]] = bitcast i[[SZ]]* [[AA_CASTED]] to i16*
// CHECK: store i16 [[AA]], i16* [[AA_C]], align
// CHECK: [[PARAM:%.+]] = load i[[SZ]], i[[SZ]]* [[AA_CASTED]], align
[OPENMP] General code improvements. llvm-svn: 330140
2018-04-17 01:59:34 +08:00
// CHECK: call {{.*}}void (%struct.ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%struct.ident_t* [[DEF_LOC]], i32 1, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, i[[SZ]])* [[OMP_OUTLINED2:@.+]] to void (i32*, i32*, ...)*), i[[SZ]] [[PARAM]])
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
//
//
// CHECK: define internal {{.*}}void [[OMP_OUTLINED2]](i32* noalias %.global_tid., i32* noalias %.bound_tid., i[[SZ]] %{{.+}})
// CHECK: [[AA_ADDR:%.+]] = alloca i[[SZ]], align
// CHECK: store i[[SZ]] %{{.+}}, i[[SZ]]* [[AA_ADDR]], align
// CHECK: [[AA_CADDR:%.+]] = bitcast i[[SZ]]* [[AA_ADDR]] to i16*
// CHECK: [[AA:%.+]] = load i16, i16* [[AA_CADDR]], align
[OPENMP]Allow cancellation constructs in target parallel regions. Summary: omp cancellation point parallel and omp cancel parallel directives are allowed in target paralle regions. Reviewers: jdoerfert Subscribers: yaxunl, guansong, caomhin, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D78941
2020-04-27 23:37:35 +08:00
// CHECK: [[IS_CANCEL:%.+]] = call i32 @__kmpc_cancel(%struct.ident_t* @{{.+}}, i32 %{{.+}}, i32 1)
// CHECK: [[CMP:%.+]] = icmp ne i32 [[IS_CANCEL]], 0
// CHECK: br i1 [[CMP]], label %[[EXIT:.+]], label %[[CONTINUE:[^,]+]]
// CHECK: [[EXIT]]:
// CHECK: br label %[[CONTINUE]]
// CHECK: [[CONTINUE]]:
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK: ret void
// CHECK-NEXT: }
// CHECK: define internal void [[HVT3]]
// Create stack storage and store argument in there.
// CHECK: [[A_ADDR:%.+]] = alloca i[[SZ]], align
// CHECK: [[AA_ADDR:%.+]] = alloca i[[SZ]], align
// CHECK: [[A_CASTED:%.+]] = alloca i[[SZ]], align
// CHECK: [[AA_CASTED:%.+]] = alloca i[[SZ]], align
// CHECK-DAG: store i[[SZ]] %{{.+}}, i[[SZ]]* [[A_ADDR]], align
// CHECK-DAG: store i[[SZ]] %{{.+}}, i[[SZ]]* [[AA_ADDR]], align
// CHECK-64-DAG:[[A_CADDR:%.+]] = bitcast i[[SZ]]* [[A_ADDR]] to i32*
// CHECK-DAG: [[AA_CADDR:%.+]] = bitcast i[[SZ]]* [[AA_ADDR]] to i16*
// CHECK-64-DAG:[[A:%.+]] = load i32, i32* [[A_CADDR]], align
// CHECK-32-DAG:[[A:%.+]] = load i32, i32* [[A_ADDR]], align
// CHECK-64-DAG:[[A_C:%.+]] = bitcast i[[SZ]]* [[A_CASTED]] to i32*
// CHECK-64-DAG:store i32 [[A]], i32* [[A_C]], align
// CHECK-32-DAG:store i32 [[A]], i32* [[A_CASTED]], align
// CHECK-DAG: [[AA:%.+]] = load i16, i16* [[AA_CADDR]], align
// CHECK-DAG: [[AA_C:%.+]] = bitcast i[[SZ]]* [[AA_CASTED]] to i16*
// CHECK-DAG: store i16 [[AA]], i16* [[AA_C]], align
// CHECK-DAG: [[PARAM1:%.+]] = load i[[SZ]], i[[SZ]]* [[A_CASTED]], align
// CHECK-DAG: [[PARAM2:%.+]] = load i[[SZ]], i[[SZ]]* [[AA_CASTED]], align
[OPENMP] General code improvements. llvm-svn: 330140
2018-04-17 01:59:34 +08:00
// CHECK-DAG: call {{.*}}void (%struct.ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%struct.ident_t* [[DEF_LOC]], i32 2, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, i[[SZ]], i[[SZ]])* [[OMP_OUTLINED3:@.+]] to void (i32*, i32*, ...)*), i[[SZ]] [[PARAM1]], i[[SZ]] [[PARAM2]])
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
//
//
// CHECK: define internal {{.*}}void [[OMP_OUTLINED3]](i32* noalias %.global_tid., i32* noalias %.bound_tid., i[[SZ]] %{{.+}}, i[[SZ]] %{{.+}})
// CHECK: [[A_ADDR:%.+]] = alloca i[[SZ]], align
// CHECK: [[AA_ADDR:%.+]] = alloca i[[SZ]], align
// CHECK-DAG: store i[[SZ]] %{{.+}}, i[[SZ]]* [[A_ADDR]], align
// CHECK-DAG: store i[[SZ]] %{{.+}}, i[[SZ]]* [[AA_ADDR]], align
// CHECK-64-DAG:[[A_CADDR:%.+]] = bitcast i[[SZ]]* [[A_ADDR]] to i32*
// CHECK-DAG: [[AA_CADDR:%.+]] = bitcast i[[SZ]]* [[AA_ADDR]] to i16*
// CHECK: ret void
// CHECK-NEXT: }
// CHECK: define internal void [[HVT4]]
// Create local storage for each capture.
// CHECK: [[LOCAL_A:%.+]] = alloca i[[SZ]]
// CHECK: [[LOCAL_B:%.+]] = alloca [10 x float]*
// CHECK: [[LOCAL_VLA1:%.+]] = alloca i[[SZ]]
// CHECK: [[LOCAL_BN:%.+]] = alloca float*
// CHECK: [[LOCAL_C:%.+]] = alloca [5 x [10 x double]]*
// CHECK: [[LOCAL_VLA2:%.+]] = alloca i[[SZ]]
// CHECK: [[LOCAL_VLA3:%.+]] = alloca i[[SZ]]
// CHECK: [[LOCAL_CN:%.+]] = alloca double*
// CHECK: [[LOCAL_D:%.+]] = alloca [[TT]]*
// CHECK: [[LOCAL_A_CASTED:%.+]] = alloca i[[SZ]]
// CHECK-DAG: store i[[SZ]] [[ARG_A:%.+]], i[[SZ]]* [[LOCAL_A]]
// CHECK-DAG: store [10 x float]* [[ARG_B:%.+]], [10 x float]** [[LOCAL_B]]
// CHECK-DAG: store i[[SZ]] [[ARG_VLA1:%.+]], i[[SZ]]* [[LOCAL_VLA1]]
// CHECK-DAG: store float* [[ARG_BN:%.+]], float** [[LOCAL_BN]]
// CHECK-DAG: store [5 x [10 x double]]* [[ARG_C:%.+]], [5 x [10 x double]]** [[LOCAL_C]]
// CHECK-DAG: store i[[SZ]] [[ARG_VLA2:%.+]], i[[SZ]]* [[LOCAL_VLA2]]
// CHECK-DAG: store i[[SZ]] [[ARG_VLA3:%.+]], i[[SZ]]* [[LOCAL_VLA3]]
// CHECK-DAG: store double* [[ARG_CN:%.+]], double** [[LOCAL_CN]]
// CHECK-DAG: store [[TT]]* [[ARG_D:%.+]], [[TT]]** [[LOCAL_D]]
// CHECK-64-DAG:[[CONV_AP:%.+]] = bitcast i[[SZ]]* [[LOCAL_A]] to i32*
// CHECK-DAG: [[REF_B:%.+]] = load [10 x float]*, [10 x float]** [[LOCAL_B]],
// CHECK-DAG: [[VAL_VLA1:%.+]] = load i[[SZ]], i[[SZ]]* [[LOCAL_VLA1]],
// CHECK-DAG: [[REF_BN:%.+]] = load float*, float** [[LOCAL_BN]],
// CHECK-DAG: [[REF_C:%.+]] = load [5 x [10 x double]]*, [5 x [10 x double]]** [[LOCAL_C]],
// CHECK-DAG: [[VAL_VLA2:%.+]] = load i[[SZ]], i[[SZ]]* [[LOCAL_VLA2]],
// CHECK-DAG: [[VAL_VLA3:%.+]] = load i[[SZ]], i[[SZ]]* [[LOCAL_VLA3]],
// CHECK-DAG: [[REF_CN:%.+]] = load double*, double** [[LOCAL_CN]],
// CHECK-DAG: [[REF_D:%.+]] = load [[TT]]*, [[TT]]** [[LOCAL_D]],
// CHECK-64-DAG:[[CONV_A:%.+]] = load i32, i32* [[CONV_AP]]
// CHECK-64-DAG:[[CONV:%.+]] = bitcast i[[SZ]]* [[LOCAL_A_CASTED]] to i32*
// CHECK-64-DAG:store i32 [[CONV_A]], i32* [[CONV]], align
// CHECK-32-DAG:[[LOCAL_AV:%.+]] = load i32, i32* [[LOCAL_A]]
// CHECK-32-DAG:store i32 [[LOCAL_AV]], i32* [[LOCAL_A_CASTED]], align
// CHECK-DAG: [[REF_A:%.+]] = load i[[SZ]], i[[SZ]]* [[LOCAL_A_CASTED]],
[OPENMP] General code improvements. llvm-svn: 330140
2018-04-17 01:59:34 +08:00
// CHECK: call {{.*}}void (%struct.ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%struct.ident_t* [[DEF_LOC]], i32 9, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, i[[SZ]], [10 x float]*, i[[SZ]], float*, [5 x [10 x double]]*, i[[SZ]], i[[SZ]], double*, [[TT]]*)* [[OMP_OUTLINED4:@.+]] to void (i32*, i32*, ...)*), i[[SZ]] [[REF_A]], [10 x float]* [[REF_B]], i[[SZ]] [[VAL_VLA1]], float* [[REF_BN]], [5 x [10 x double]]* [[REF_C]], i[[SZ]] [[VAL_VLA2]], i[[SZ]] [[VAL_VLA3]], double* [[REF_CN]], [[TT]]* [[REF_D]])
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
//
//
[OPENMP] Fix PR35013: Fix passing VLAs captures to outlined functions. Fixed passing of VLAs and variably-modified types to outlined functions. Synchronized passing with the types codegen. llvm-svn: 316488
2017-10-25 03:52:31 +08:00
// CHECK: define internal {{.*}}void [[OMP_OUTLINED4]](i32* noalias %.global_tid., i32* noalias %.bound_tid., i[[SZ]] %{{.+}}, [10 x float]* {{.+}}, i[[SZ]] %{{.+}}, float* {{.+}}, [5 x [10 x double]]* {{.+}}, i[[SZ]] %{{.+}}, i[[SZ]] %{{.+}}, double* {{.+}}, [[TT]]* {{.+}})
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// To reduce complexity, we're only going as far as validating the signature of the outlined parallel function.
template<typename tx>
tx ftemplate(int n) {
tx a = 0;
short aa = 0;
tx b[10];
#pragma omp target parallel if(target: n>40)
{
a += 1;
aa += 1;
b[2] += 1;
}
return a;
}
static
int fstatic(int n) {
int a = 0;
short aa = 0;
char aaa = 0;
int b[10];
#pragma omp target parallel if(target: n>50)
{
a += 1;
aa += 1;
aaa += 1;
b[2] += 1;
}
return a;
}
struct S1 {
double a;
int r1(int n){
int b = n+1;
short int c[2][n];
#pragma omp target parallel if(target: n>60)
{
this->a = (double)b + 1.5;
c[1][1] = ++a;
}
return c[1][1] + (int)b;
}
};
// CHECK: define {{.*}}@{{.*}}bar{{.*}}
int bar(int n){
int a = 0;
// CHECK: call {{.*}}i32 [[FOO]](i32 {{.*}})
a += foo(n);
S1 S;
// CHECK: call {{.*}}i32 [[FS1:@.+]]([[S1]]* {{.*}}, i32 {{.*}})
a += S.r1(n);
// CHECK: call {{.*}}i32 [[FSTATIC:@.+]](i32 {{.*}})
a += fstatic(n);
// CHECK: call {{.*}}i32 [[FTEMPLATE:@.+]](i32 {{.*}})
a += ftemplate<int>(n);
return a;
}
//
// CHECK: define {{.*}}[[FS1]]
//
// CHECK: i8* @llvm.stacksave()
// CHECK-64: [[B_ADDR:%.+]] = bitcast i[[SZ]]* [[B_CADDR:%.+]] to i32*
// CHECK-64: store i32 %{{.+}}, i32* [[B_ADDR]],
// CHECK-64: [[B_CVAL:%.+]] = load i[[SZ]], i[[SZ]]* [[B_CADDR]],
[DebugInfo] Avoid name conflict of generated VLA expression variable. Summary: This patch also adds the 'DW_AT_artificial' flag to the generated variable. Addresses the issues mentioned in http://llvm.org/PR30553. Reviewers: CarlosAlbertoEnciso, probinson, aprantl Reviewed By: aprantl Subscribers: JDevlieghere, cfe-commits Differential Revision: https://reviews.llvm.org/D43189 llvm-svn: 324988
2018-02-13 15:49:34 +08:00
// CHECK-32: store i32 %{{.+}}, i32* %__vla_expr
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-32: store i32 %{{.+}}, i32* [[B_ADDR:%.+]],
// CHECK-32: [[B_CVAL:%.+]] = load i[[SZ]], i[[SZ]]* [[B_ADDR]],
[OPENMP] Add codegen for `depend` clauses on `target` directive. Added basic support for codegen of `depend` clauses on `target` directive. llvm-svn: 322501
2018-01-16 03:06:12 +08:00
// CHECK: [[IF:%.+]] = icmp sgt i32 {{[^,]+}}, 60
// CHECK: br i1 [[IF]], label %[[TRY:[^,]+]], label %[[FAIL:[^,]+]]
// CHECK: [[TRY]]
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// We capture 2 VLA sizes in this target region
// CHECK: [[CELEMSIZE2:%.+]] = mul nuw i[[SZ]] 2, [[VLA0:%.+]]
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-64: [[CSIZE:%.+]] = mul nuw i64 [[CELEMSIZE2]], 2
// CHECK-32: [[CSZSIZE:%.+]] = mul nuw i32 [[CELEMSIZE2]], 2
// CHECK-32: [[CSIZE:%.+]] = sext i32 [[CSZSIZE]] to i64
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OpenMP] Add Location Fields to Libomptarget Runtime for Debugging Summary: Add support for passing source locations to libomptarget runtime functions using the ident_t struct present in the rest of the libomp API. This will allow the runtime system to give much more insightful error messages and debugging values. Reviewers: jdoerfert grokos Differential Revision: https://reviews.llvm.org/D87946
2020-11-20 00:56:59 +08:00
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(%struct.ident_t* @{{.+}}, i64 -1, i8* @{{[^,]+}}, i32 5, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* [[SR:%[^,]+]], i64* getelementptr inbounds ([5 x i64], [5 x i64]* [[MAPT7]], i32 0, i32 0), i8** null, i8** null, i32 1, i32 0)
[OPENMP]Fix PR48076: mapping of data member pointer. If the data member pointer is mapped, the compiler tries to optimize the mapping of such data by discarding explicit mapping flags and trying to emit combined data instead. In some cases, this optimization is not quite correctly implemented and it leads to a program crash at the runtime. Instead, if the data member is mapped, just emit it as is and do not emit combined mapping flags for it. Differential Revision: https://reviews.llvm.org/D91552
2020-11-17 01:43:53 +08:00
// CHECK-DAG: [[BPR]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP:%.+]], i32 0, i32 0
// CHECK-DAG: [[PR]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P:%.+]], i32 0, i32 0
// CHECK-DAG: [[SR]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S:%.+]], i32 0, i32 0
// CHECK-DAG: [[SADDR0:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX0:[0-9]+]]
// CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX0]]
// CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX0]]
// CHECK-DAG: [[SADDR1:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX1:[0-9]+]]
// CHECK-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX1]]
// CHECK-DAG: [[PADDR1:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX1]]
// CHECK-DAG: [[SADDR2:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX2:[0-9]+]]
// CHECK-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX2]]
// CHECK-DAG: [[PADDR2:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX2]]
// CHECK-DAG: [[SADDR3:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX3:[0-9]+]]
// CHECK-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX3]]
// CHECK-DAG: [[PADDR3:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX3]]
// CHECK-DAG: [[SADDR4:%.+]] = getelementptr inbounds [5 x i64], [5 x i64]* [[S]], i32 [[IDX4:[0-9]+]]
// CHECK-DAG: [[BPADDR4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[BP]], i32 [[IDX4]]
// CHECK-DAG: [[PADDR4:%.+]] = getelementptr inbounds [5 x i8*], [5 x i8*]* [[P]], i32 [[IDX4]]
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// The names below are not necessarily consistent with the names used for the
// addresses above as some are repeated.
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CBPADDR0:%.+]],
// CHECK-DAG: store i[[SZ]] [[VLA0]], i[[SZ]]* [[CPADDR0:%.+]],
// CHECK-DAG: [[CBPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR0]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CBPADDR1:%.+]],
// CHECK-DAG: store i[[SZ]] 2, i[[SZ]]* [[CPADDR1:%.+]],
// CHECK-DAG: [[CBPADDR1]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR1]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CBPADDR2:%.+]],
// CHECK-DAG: store i[[SZ]] [[B_CVAL]], i[[SZ]]* [[CPADDR2:%.+]],
// CHECK-DAG: [[CBPADDR2]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
// CHECK-DAG: [[CPADDR2]] = bitcast i8** {{%[^,]+}} to i[[SZ]]*
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: store [[S1]]* %{{.+}}, [[S1]]** [[CBPADDR3:%.+]],
[OPENMP]Fix PR48076: mapping of data member pointer. If the data member pointer is mapped, the compiler tries to optimize the mapping of such data by discarding explicit mapping flags and trying to emit combined data instead. In some cases, this optimization is not quite correctly implemented and it leads to a program crash at the runtime. Instead, if the data member is mapped, just emit it as is and do not emit combined mapping flags for it. Differential Revision: https://reviews.llvm.org/D91552
2020-11-17 01:43:53 +08:00
// CHECK-DAG: store double* %{{.+}}, double** [[CPADDR3:%.+]],
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: [[CBPADDR3]] = bitcast i8** {{%[^,]+}} to [[S1]]**
[OPENMP]Fix PR48076: mapping of data member pointer. If the data member pointer is mapped, the compiler tries to optimize the mapping of such data by discarding explicit mapping flags and trying to emit combined data instead. In some cases, this optimization is not quite correctly implemented and it leads to a program crash at the runtime. Instead, if the data member is mapped, just emit it as is and do not emit combined mapping flags for it. Differential Revision: https://reviews.llvm.org/D91552
2020-11-17 01:43:53 +08:00
// CHECK-DAG: [[CPADDR3]] = bitcast i8** {{%[^,]+}} to double*
The patch adds support for the new map interface between clang and libomptarget. The changes in the interface are the following: device IDs are now 64-bit integers (as opposed to 32-bit) map flags are 64-bit long (used to be 32-bit) mappings for partially mapped structs are now calculated at compile time and members of partially mapped structs are flagged using the MEMBER_OF field Support for is_device_ptr on struct members was dropped - this functionality is not supported by the OpenMP standard and its implementation is technically infeasible (however, use_device_ptr on struct members works as a non-standard extension of the compiler) llvm-svn: 337468
2018-07-20 00:34:13 +08:00
[OpenMP] Update CHECK-DAG usage in target_parallel_codegen.cpp This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. There were two issues in this test: 1. There were sets of patterns for store instructions in which a pattern X could match a superset of a pattern Y. While X appeared before Y, Y's intended match appeared before X's intended match. The result was that X matched Y's intended match. Under the old overlapping behavior, Y also matched Y's intended match. Under the new non-overlapping behavior, Y had nothing left to match. This patch fixes this by gathering these sets in one place and putting the most specific patterns (Y) before the more general patterns (X). 2. The CHECK-DAG patterns involving the variables CBPADDR3 and CBPADDR4 were the same, but there was only one match in the text, so CBPADDR4 patterns had nothing to match under the new non-overlapping behavior. Moreover, a preceding related series of directives had variables (SADDR0, BPADDR0, etc.) numbered only 0 through 4, but this series had variables numbered 0 through 5. Assuming CBPADDR4's directives were not intended, this patch removes them. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54765 llvm-svn: 347351
2018-11-21 06:05:23 +08:00
// CHECK-DAG: store i16* %{{.+}}, i16** [[CBPADDR4:%.+]],
// CHECK-DAG: store i16* %{{.+}}, i16** [[CPADDR4:%.+]],
// CHECK-DAG: [[CBPADDR4]] = bitcast i8** {{%[^,]+}} to i16**
// CHECK-DAG: [[CPADDR4]] = bitcast i8** {{%[^,]+}} to i16**
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP]Fix PR41966: type mismatch in runtime functions. Target-based runtime functions use int64_t type for sizes, while the compiler uses size_t type. It leads to miscompilation in 32 bit mode. llvm-svn: 364327
2019-06-26 00:00:43 +08:00
// CHECK-DAG: store i64 4, i64* {{%[^,]+}}
// CHECK-DAG: store i64 [[CSIZE]], i64* {{%[^,]+}}
// CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}}
// CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}}
[OPENMP]Fix PR48076: mapping of data member pointer. If the data member pointer is mapped, the compiler tries to optimize the mapping of such data by discarding explicit mapping flags and trying to emit combined data instead. In some cases, this optimization is not quite correctly implemented and it leads to a program crash at the runtime. Instead, if the data member is mapped, just emit it as is and do not emit combined mapping flags for it. Differential Revision: https://reviews.llvm.org/D91552
2020-11-17 01:43:53 +08:00
// CHECK-DAG: store i64 {{4|8}}, i64* {{%[^,]+}}
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP] Simplify codegen for non-offloading code. Simplified and generalized codegen for non-offloading part that works if offloading is failed or condition of the `if` clause is `false`. llvm-svn: 314670
2017-10-02 22:20:58 +08:00
// CHECK-NEXT: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:[^,]+]], label %[[END:[^,]+]]
// CHECK: [[FAIL]]
// CHECK: call void [[HVT7:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}})
// CHECK-NEXT: br label %[[END]]
// CHECK: [[END]]
//
// CHECK: define {{.*}}[[FSTATIC]]
//
// CHECK: [[IF:%.+]] = icmp sgt i32 {{[^,]+}}, 50
// CHECK: br i1 [[IF]], label %[[IFTHEN:[^,]+]], label %[[IFELSE:[^,]+]]
// CHECK: [[IFTHEN]]
[OpenMP] Add Location Fields to Libomptarget Runtime for Debugging Summary: Add support for passing source locations to libomptarget runtime functions using the ident_t struct present in the rest of the libomp API. This will allow the runtime system to give much more insightful error messages and debugging values. Reviewers: jdoerfert grokos Differential Revision: https://reviews.llvm.org/D87946
2020-11-20 00:56:59 +08:00
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(%struct.ident_t* @{{.+}}, i64 -1, i8* @{{[^,]+}}, i32 4, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* getelementptr inbounds ([4 x i64], [4 x i64]* [[SIZET6]], i32 0, i32 0), i64* getelementptr inbounds ([4 x i64], [4 x i64]* [[MAPT6]], i32 0, i32 0), i8** null, i8** null, i32 1, i32 0)
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPR]] = getelementptr inbounds [4 x i8*], [4 x i8*]* [[BP:%.+]], i32 0, i32 0
// CHECK-DAG: [[PR]] = getelementptr inbounds [4 x i8*], [4 x i8*]* [[P:%.+]], i32 0, i32 0
// CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [4 x i8*], [4 x i8*]* [[BP]], i32 0, i32 0
// CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [4 x i8*], [4 x i8*]* [[P]], i32 0, i32 0
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: [[CBPADDR0:%.+]] = bitcast i8** [[BPADDR0]] to i[[SZ]]*
// CHECK-DAG: [[CPADDR0:%.+]] = bitcast i8** [[PADDR0]] to i[[SZ]]*
// CHECK-DAG: store i[[SZ]] [[VAL0:%.+]], i[[SZ]]* [[CBPADDR0]],
// CHECK-DAG: store i[[SZ]] [[VAL0]], i[[SZ]]* [[CPADDR0]],
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [4 x i8*], [4 x i8*]* [[BP]], i32 0, i32 1
// CHECK-DAG: [[PADDR1:%.+]] = getelementptr inbounds [4 x i8*], [4 x i8*]* [[P]], i32 0, i32 1
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: [[CBPADDR1:%.+]] = bitcast i8** [[BPADDR1]] to i[[SZ]]*
// CHECK-DAG: [[CPADDR1:%.+]] = bitcast i8** [[PADDR1]] to i[[SZ]]*
// CHECK-DAG: store i[[SZ]] [[VAL1:%.+]], i[[SZ]]* [[CBPADDR1]],
// CHECK-DAG: store i[[SZ]] [[VAL1]], i[[SZ]]* [[CPADDR1]],
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [4 x i8*], [4 x i8*]* [[BP]], i32 0, i32 2
// CHECK-DAG: [[PADDR2:%.+]] = getelementptr inbounds [4 x i8*], [4 x i8*]* [[P]], i32 0, i32 2
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: [[CBPADDR2:%.+]] = bitcast i8** [[BPADDR2]] to i[[SZ]]*
// CHECK-DAG: [[CPADDR2:%.+]] = bitcast i8** [[PADDR2]] to i[[SZ]]*
// CHECK-DAG: store i[[SZ]] [[VAL2:%.+]], i[[SZ]]* [[CBPADDR2]],
// CHECK-DAG: store i[[SZ]] [[VAL2]], i[[SZ]]* [[CPADDR2]],
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPADDR3:%.+]] = getelementptr inbounds [4 x i8*], [4 x i8*]* [[BP]], i32 0, i32 3
// CHECK-DAG: [[PADDR3:%.+]] = getelementptr inbounds [4 x i8*], [4 x i8*]* [[P]], i32 0, i32 3
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: [[CBPADDR3:%.+]] = bitcast i8** [[BPADDR3]] to [10 x i32]**
// CHECK-DAG: [[CPADDR3:%.+]] = bitcast i8** [[PADDR3]] to [10 x i32]**
// CHECK-DAG: store [10 x i32]* [[VAL3:%.+]], [10 x i32]** [[CBPADDR3]],
// CHECK-DAG: store [10 x i32]* [[VAL3]], [10 x i32]** [[CPADDR3]],
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP] Simplify codegen for non-offloading code. Simplified and generalized codegen for non-offloading part that works if offloading is failed or condition of the `if` clause is `false`. llvm-svn: 314670
2017-10-02 22:20:58 +08:00
// CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
// CHECK: [[FAIL]]
// CHECK: call void [[HVT6:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}})
// CHECK-NEXT: br label %[[END]]
// CHECK: [[END]]
[OPENMP] Simplify codegen for non-offloading code. Simplified and generalized codegen for non-offloading part that works if offloading is failed or condition of the `if` clause is `false`. llvm-svn: 314670
2017-10-02 22:20:58 +08:00
// CHECK-NEXT: br label %[[IFEND:.+]]
// CHECK: [[IFELSE]]
// CHECK: call void [[HVT6]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}}, {{[^,]+}})
// CHECK-NEXT: br label %[[IFEND]]
// CHECK: [[IFEND]]
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
//
// CHECK: define {{.*}}[[FTEMPLATE]]
//
// CHECK: [[IF:%.+]] = icmp sgt i32 {{[^,]+}}, 40
// CHECK: br i1 [[IF]], label %[[IFTHEN:[^,]+]], label %[[IFELSE:[^,]+]]
// CHECK: [[IFTHEN]]
[OpenMP] Add Location Fields to Libomptarget Runtime for Debugging Summary: Add support for passing source locations to libomptarget runtime functions using the ident_t struct present in the rest of the libomp API. This will allow the runtime system to give much more insightful error messages and debugging values. Reviewers: jdoerfert grokos Differential Revision: https://reviews.llvm.org/D87946
2020-11-20 00:56:59 +08:00
// CHECK-DAG: [[RET:%.+]] = call i32 @__tgt_target_teams_mapper(%struct.ident_t* @{{.+}}, i64 -1, i8* @{{[^,]+}}, i32 3, i8** [[BPR:%[^,]+]], i8** [[PR:%[^,]+]], i64* getelementptr inbounds ([3 x i64], [3 x i64]* [[SIZET5]], i32 0, i32 0), i64* getelementptr inbounds ([3 x i64], [3 x i64]* [[MAPT5]], i32 0, i32 0), i8** null, i8** null, i32 1, i32 0)
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPR]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[BP:%.+]], i32 0, i32 0
// CHECK-DAG: [[PR]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[P:%.+]], i32 0, i32 0
// CHECK-DAG: [[BPADDR0:%.+]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[BP]], i32 0, i32 0
// CHECK-DAG: [[PADDR0:%.+]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[P]], i32 0, i32 0
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: [[CBPADDR0:%.+]] = bitcast i8** [[BPADDR0]] to i[[SZ]]*
// CHECK-DAG: [[CPADDR0:%.+]] = bitcast i8** [[PADDR0]] to i[[SZ]]*
// CHECK-DAG: store i[[SZ]] [[VAL0:%.+]], i[[SZ]]* [[CBPADDR0]],
// CHECK-DAG: store i[[SZ]] [[VAL0]], i[[SZ]]* [[CPADDR0]],
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPADDR1:%.+]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[BP]], i32 0, i32 1
// CHECK-DAG: [[PADDR1:%.+]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[P]], i32 0, i32 1
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: [[CBPADDR1:%.+]] = bitcast i8** [[BPADDR1]] to i[[SZ]]*
// CHECK-DAG: [[CPADDR1:%.+]] = bitcast i8** [[PADDR1]] to i[[SZ]]*
// CHECK-DAG: store i[[SZ]] [[VAL1:%.+]], i[[SZ]]* [[CBPADDR1]],
// CHECK-DAG: store i[[SZ]] [[VAL1]], i[[SZ]]* [[CPADDR1]],
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-DAG: [[BPADDR2:%.+]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[BP]], i32 0, i32 2
// CHECK-DAG: [[PADDR2:%.+]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[P]], i32 0, i32 2
[OPENMP][DEBUG] Generate second function with correct arg types. Currently, if the some of the parameters are captured by value, this argument is converted to uintptr_t type and thus we loosing the debug info about real type of the argument (captured variable): ``` void @.outlined_function.(uintptr %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.outlined_function.(uintptr %a.casted.val) ... ``` To resolve this problem, in debug mode a speciall external wrapper function is generated, that calls the outlined function with the correct parameters types: ``` void @.wrapper.(uintptr %par) { %a = alloca i32 %cast = bitcast i32* %a to uintptr* store uintptr %par, uintptr *%cast %a.val = load i32, i32* %a call void @.outlined_function.(i32 %a) ret void } void @.outlined_function.(i32 %par); ... %a = alloca i32 %a.casted = alloca uintptr %cast = bitcast uintptr* %a.casted to i32* %a.val = load i32, i32 *%a store i32 %a.val, i32 *%cast %a.casted.val = load uintptr, uintptr* %a.casted call void @.wrapper.(uintptr %a.casted.val) ... ``` llvm-svn: 306697
2017-06-30 00:43:05 +08:00
// CHECK-DAG: [[CBPADDR2:%.+]] = bitcast i8** [[BPADDR2]] to [10 x i32]**
// CHECK-DAG: [[CPADDR2:%.+]] = bitcast i8** [[PADDR2]] to [10 x i32]**
// CHECK-DAG: store [10 x i32]* [[VAL2:%.+]], [10 x i32]** [[CBPADDR2]],
// CHECK-DAG: store [10 x i32]* [[VAL2]], [10 x i32]** [[CPADDR2]],
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
[OPENMP] Simplify codegen for non-offloading code. Simplified and generalized codegen for non-offloading part that works if offloading is failed or condition of the `if` clause is `false`. llvm-svn: 314670
2017-10-02 22:20:58 +08:00
// CHECK: [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// CHECK-NEXT: br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
// CHECK: [[FAIL]]
// CHECK: call void [[HVT5:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}})
// CHECK-NEXT: br label %[[END]]
// CHECK: [[END]]
[OPENMP] Simplify codegen for non-offloading code. Simplified and generalized codegen for non-offloading part that works if offloading is failed or condition of the `if` clause is `false`. llvm-svn: 314670
2017-10-02 22:20:58 +08:00
// CHECK-NEXT: br label %[[IFEND:.+]]
// CHECK: [[IFELSE]]
// CHECK: call void [[HVT:@.+]]({{[^,]+}}, {{[^,]+}}, {{[^,]+}})
// CHECK-NEXT: br label %[[IFEND]]
// CHECK: [[IFEND]]
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// Check that the offloading functions are emitted and that the arguments are
// correct and loaded correctly for the target regions of the callees of bar().
// CHECK: define internal void [[HVT7]]
// Create local storage for each capture.
// CHECK: [[LOCAL_THIS:%.+]] = alloca [[S1]]*
// CHECK: [[LOCAL_B:%.+]] = alloca i[[SZ]]
// CHECK: [[LOCAL_VLA1:%.+]] = alloca i[[SZ]]
// CHECK: [[LOCAL_VLA2:%.+]] = alloca i[[SZ]]
// CHECK: [[LOCAL_C:%.+]] = alloca i16*
// CHECK: [[LOCAL_B_CASTED:%.+]] = alloca i[[SZ]]
// CHECK-DAG: store [[S1]]* [[ARG_THIS:%.+]], [[S1]]** [[LOCAL_THIS]]
// CHECK-DAG: store i[[SZ]] [[ARG_B:%.+]], i[[SZ]]* [[LOCAL_B]]
// CHECK-DAG: store i[[SZ]] [[ARG_VLA1:%.+]], i[[SZ]]* [[LOCAL_VLA1]]
// CHECK-DAG: store i[[SZ]] [[ARG_VLA2:%.+]], i[[SZ]]* [[LOCAL_VLA2]]
// CHECK-DAG: store i16* [[ARG_C:%.+]], i16** [[LOCAL_C]]
// Store captures in the context.
// CHECK-DAG: [[REF_THIS:%.+]] = load [[S1]]*, [[S1]]** [[LOCAL_THIS]],
// CHECK-64-DAG:[[CONV_BP:%.+]] = bitcast i[[SZ]]* [[LOCAL_B]] to i32*
// CHECK-DAG: [[VAL_VLA1:%.+]] = load i[[SZ]], i[[SZ]]* [[LOCAL_VLA1]],
// CHECK-DAG: [[VAL_VLA2:%.+]] = load i[[SZ]], i[[SZ]]* [[LOCAL_VLA2]],
// CHECK-DAG: [[REF_C:%.+]] = load i16*, i16** [[LOCAL_C]],
// CHECK-64-DAG:[[CONV_B:%.+]] = load i32, i32* [[CONV_BP]]
// CHECK-64-DAG:[[CONV:%.+]] = bitcast i[[SZ]]* [[LOCAL_B_CASTED]] to i32*
// CHECK-64-DAG:store i32 [[CONV_B]], i32* [[CONV]], align
// CHECK-32-DAG:[[LOCAL_BV:%.+]] = load i32, i32* [[LOCAL_B]]
// CHECK-32-DAG:store i32 [[LOCAL_BV]], i32* [[LOCAL_B_CASTED]], align
// CHECK-DAG: [[REF_B:%.+]] = load i[[SZ]], i[[SZ]]* [[LOCAL_B_CASTED]],
[OPENMP] General code improvements. llvm-svn: 330140
2018-04-17 01:59:34 +08:00
// CHECK: call {{.*}}void (%struct.ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%struct.ident_t* [[DEF_LOC]], i32 5, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, [[S1]]*, i[[SZ]], i[[SZ]], i[[SZ]], i16*)* [[OMP_OUTLINED5:@.+]] to void (i32*, i32*, ...)*), [[S1]]* [[REF_THIS]], i[[SZ]] [[REF_B]], i[[SZ]] [[VAL_VLA1]], i[[SZ]] [[VAL_VLA2]], i16* [[REF_C]])
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
//
//
[OPENMP] Fix PR35013: Fix passing VLAs captures to outlined functions. Fixed passing of VLAs and variably-modified types to outlined functions. Synchronized passing with the types codegen. llvm-svn: 316488
2017-10-25 03:52:31 +08:00
// CHECK: define internal {{.*}}void [[OMP_OUTLINED5]](i32* noalias %.global_tid., i32* noalias %.bound_tid., [[S1]]* %{{.+}}, i[[SZ]] %{{.+}}, i[[SZ]] %{{.+}}, i[[SZ]] %{{.+}}, i16* {{.+}})
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
// To reduce complexity, we're only going as far as validating the signature of the outlined parallel function.
// CHECK: define internal void [[HVT6]]
// Create local storage for each capture.
// CHECK: [[LOCAL_A:%.+]] = alloca i[[SZ]]
// CHECK: [[LOCAL_AA:%.+]] = alloca i[[SZ]]
// CHECK: [[LOCAL_AAA:%.+]] = alloca i[[SZ]]
// CHECK: [[LOCAL_B:%.+]] = alloca [10 x i32]*
// CHECK: [[LOCAL_A_CASTED:%.+]] = alloca i[[SZ]]
// CHECK: [[LOCAL_AA_CASTED:%.+]] = alloca i[[SZ]]
// CHECK: [[LOCAL_AAA_CASTED:%.+]] = alloca i[[SZ]]
// CHECK-DAG: store i[[SZ]] [[ARG_A:%.+]], i[[SZ]]* [[LOCAL_A]]
// CHECK-DAG: store i[[SZ]] [[ARG_AA:%.+]], i[[SZ]]* [[LOCAL_AA]]
// CHECK-DAG: store i[[SZ]] [[ARG_AAA:%.+]], i[[SZ]]* [[LOCAL_AAA]]
// CHECK-DAG: store [10 x i32]* [[ARG_B:%.+]], [10 x i32]** [[LOCAL_B]]
// Store captures in the context.
// CHECK-64-DAG:[[CONV_AP:%.+]] = bitcast i[[SZ]]* [[LOCAL_A]] to i32*
// CHECK-DAG: [[CONV_AAP:%.+]] = bitcast i[[SZ]]* [[LOCAL_AA]] to i16*
// CHECK-DAG: [[CONV_AAAP:%.+]] = bitcast i[[SZ]]* [[LOCAL_AAA]] to i8*
// CHECK-DAG: [[REF_B:%.+]] = load [10 x i32]*, [10 x i32]** [[LOCAL_B]],
// CHECK-64-DAG:[[CONV_A:%.+]] = load i32, i32* [[CONV_AP]]
// CHECK-64-DAG:[[CONV:%.+]] = bitcast i[[SZ]]* [[LOCAL_A_CASTED]] to i32*
// CHECK-64-DAG:store i32 [[CONV_A]], i32* [[CONV]], align
// CHECK-32-DAG:[[LOCAL_AV:%.+]] = load i32, i32* [[LOCAL_A]]
// CHECK-32-DAG:store i32 [[LOCAL_AV]], i32* [[LOCAL_A_CASTED]], align
// CHECK-DAG: [[REF_A:%.+]] = load i[[SZ]], i[[SZ]]* [[LOCAL_A_CASTED]],
// CHECK-DAG: [[CONV_AA:%.+]] = load i16, i16* [[CONV_AAP]]
// CHECK-DAG: [[CONV:%.+]] = bitcast i[[SZ]]* [[LOCAL_AA_CASTED]] to i16*
// CHECK-DAG: store i16 [[CONV_AA]], i16* [[CONV]], align
// CHECK-DAG: [[REF_AA:%.+]] = load i[[SZ]], i[[SZ]]* [[LOCAL_AA_CASTED]],
// CHECK-DAG: [[CONV_AAA:%.+]] = load i8, i8* [[CONV_AAAP]]
// CHECK-DAG: [[CONV:%.+]] = bitcast i[[SZ]]* [[LOCAL_AAA_CASTED]] to i8*
// CHECK-DAG: store i8 [[CONV_AAA]], i8* [[CONV]], align
// CHECK-DAG: [[REF_AAA:%.+]] = load i[[SZ]], i[[SZ]]* [[LOCAL_AAA_CASTED]],
[OPENMP] General code improvements. llvm-svn: 330140
2018-04-17 01:59:34 +08:00
// CHECK: call {{.*}}void (%struct.ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%struct.ident_t* [[DEF_LOC]], i32 4, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, i[[SZ]], i[[SZ]], i[[SZ]], [10 x i32]*)* [[OMP_OUTLINED6:@.+]] to void (i32*, i32*, ...)*), i[[SZ]] [[REF_A]], i[[SZ]] [[REF_AA]], i[[SZ]] [[REF_AAA]], [10 x i32]* [[REF_B]])
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
//
//
// CHECK: define internal {{.*}}void [[OMP_OUTLINED6]](i32* noalias %.global_tid., i32* noalias %.bound_tid., i[[SZ]] %{{.+}}, i[[SZ]] %{{.+}}, i[[SZ]] %{{.+}}, [10 x i32]* {{.+}})
// To reduce complexity, we're only going as far as validating the signature of the outlined parallel function.
// CHECK: define internal void [[HVT5]]
// Create local storage for each capture.
// CHECK: [[LOCAL_A:%.+]] = alloca i[[SZ]]
// CHECK: [[LOCAL_AA:%.+]] = alloca i[[SZ]]
// CHECK: [[LOCAL_B:%.+]] = alloca [10 x i32]*
// CHECK: [[LOCAL_A_CASTED:%.+]] = alloca i[[SZ]]
// CHECK: [[LOCAL_AA_CASTED:%.+]] = alloca i[[SZ]]
// CHECK-DAG: store i[[SZ]] [[ARG_A:%.+]], i[[SZ]]* [[LOCAL_A]]
// CHECK-DAG: store i[[SZ]] [[ARG_AA:%.+]], i[[SZ]]* [[LOCAL_AA]]
// CHECK-DAG: store [10 x i32]* [[ARG_B:%.+]], [10 x i32]** [[LOCAL_B]]
// Store captures in the context.
// CHECK-64-DAG:[[CONV_AP:%.+]] = bitcast i[[SZ]]* [[LOCAL_A]] to i32*
// CHECK-DAG: [[CONV_AAP:%.+]] = bitcast i[[SZ]]* [[LOCAL_AA]] to i16*
// CHECK-DAG: [[REF_B:%.+]] = load [10 x i32]*, [10 x i32]** [[LOCAL_B]],
// CHECK-64-DAG:[[CONV_A:%.+]] = load i32, i32* [[CONV_AP]]
// CHECK-64-DAG:[[CONV:%.+]] = bitcast i[[SZ]]* [[LOCAL_A_CASTED]] to i32*
// CHECK-64-DAG:store i32 [[CONV_A]], i32* [[CONV]], align
// CHECK-32-DAG:[[LOCAL_AV:%.+]] = load i32, i32* [[LOCAL_A]]
// CHECK-32-DAG:store i32 [[LOCAL_AV]], i32* [[LOCAL_A_CASTED]], align
// CHECK-DAG: [[REF_A:%.+]] = load i[[SZ]], i[[SZ]]* [[LOCAL_A_CASTED]],
// CHECK-DAG: [[CONV_AA:%.+]] = load i16, i16* [[CONV_AAP]]
// CHECK-DAG: [[CONV:%.+]] = bitcast i[[SZ]]* [[LOCAL_AA_CASTED]] to i16*
// CHECK-DAG: store i16 [[CONV_AA]], i16* [[CONV]], align
// CHECK-DAG: [[REF_AA:%.+]] = load i[[SZ]], i[[SZ]]* [[LOCAL_AA_CASTED]],
[OPENMP] General code improvements. llvm-svn: 330140
2018-04-17 01:59:34 +08:00
// CHECK: call {{.*}}void (%struct.ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%struct.ident_t* [[DEF_LOC]], i32 3, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, i[[SZ]], i[[SZ]], [10 x i32]*)* [[OMP_OUTLINED7:@.+]] to void (i32*, i32*, ...)*), i[[SZ]] [[REF_A]], i[[SZ]] [[REF_AA]], [10 x i32]* [[REF_B]])
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
//
//
// CHECK: define internal {{.*}}void [[OMP_OUTLINED7]](i32* noalias %.global_tid., i32* noalias %.bound_tid., i[[SZ]] %{{.+}}, i[[SZ]] %{{.+}}, [10 x i32]* {{.+}})
// To reduce complexity, we're only going as far as validating the signature of the outlined parallel function.
[OpenMP] Add support for registering requires directives with the runtime Summary: This patch adds support for the registration of the requires directives with the runtime. Each requires directive clause will enable a particular flag to be set. The set of flags is passed to the runtime to be checked for compatibility with other such flags coming from other object files. The registration function is called whenever OpenMP is present even if a requires directive is not present. This helps detect cases in which requires directives are used inconsistently. Reviewers: ABataev, AlexEichenberger, caomhin Reviewed By: ABataev, AlexEichenberger Subscribers: jholewinski, guansong, jfb, jdoerfert, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D60568 llvm-svn: 361298
2019-05-22 03:42:01 +08:00
// CHECK: define internal void @.omp_offloading.requires_reg()
// CHECK: call void @__tgt_register_requires(i64 1)
// CHECK: ret void
[OpenMP] Codegen support for 'target parallel' on the host. This patch adds support for codegen of 'target parallel' on the host. It is also the first combined directive that requires two or more captured statements. Support for this functionality is included in the patch. A combined directive such as 'target parallel' has two captured statements, one for the 'target' and the other for the 'parallel' region. Two captured statements are required because each has different implicit parameters (see SemaOpenMP.cpp). For example, the 'parallel' has 'global_tid' and 'bound_tid' while the 'target' does not. The patch adds support for handling multiple captured statements based on the combined directive. When codegen'ing the 'target parallel' directive, the 'target' outlined function is created using the outer captured statement and the 'parallel' outlined function is created using the inner captured statement. Reviewers: ABataev Differential Revision: https://reviews.llvm.org/D28753 llvm-svn: 292419
2017-01-19 02:18:53 +08:00
#endif