Skip to content

Commit 4db88a5

Browse files
saiislamsaiislam
committed
[OpenMP][Clang] Move partial support of reverse offload to a future version
OpenMP Spec 5.2 requires unimplemented requires clauses to produce compile time error termination. Moving current partial support of reverse_offload to a distant future version 9.9 so that existing code can be tested and maintained until a complete implementation is available. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D119256
1 parent 99580e2 commit 4db88a5

File tree

6 files changed

+74
-19
lines changed

6 files changed

+74
-19
lines changed

‎clang/test/OpenMP/requires_ast_print.cpp‎

Lines changed: 11 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -5,6 +5,14 @@
55
// RUN: %clang_cc1 -verify -fopenmp-simd -ast-print %s | FileCheck %s
66
// RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -emit-pch -o %t %s
77
// RUN: %clang_cc1 -fopenmp-simd -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s
8+
9+
// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=99 -DOMP99 -ast-print %s | FileCheck --check-prefixes=CHECK,REV %s
10+
// RUN: %clang_cc1 -fopenmp -fopenmp-version=99 -DOMP99 -x c++ -std=c++11 -emit-pch -o %t %s
11+
// RUN: %clang_cc1 -fopenmp -fopenmp-version=99 -DOMP99 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck --check-prefixes=CHECK,REV %s
12+
13+
// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=99 -DOMP99 -ast-print %s | FileCheck --check-prefixes=CHECK,REV %s
14+
// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=99 -DOMP99 -x c++ -std=c++11 -emit-pch -o %t %s
15+
// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=99 -DOMP99 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck --check-prefixes=CHECK,REV %s
816
// expected-no-diagnostics
917

1018
#ifndef HEADER
@@ -16,8 +24,10 @@
1624
#pragma omp requires unified_shared_memory
1725
// CHECK:#pragma omp requires unified_shared_memory
1826

27+
#ifdef OMP99
1928
#pragma omp requires reverse_offload
20-
// CHECK:#pragma omp requires reverse_offload
29+
// REV:#pragma omp requires reverse_offload
30+
#endif
2131

2232
#pragma omp requires dynamic_allocators
2333
// CHECK:#pragma omp requires dynamic_allocators

‎clang/test/OpenMP/requires_messages.cpp‎

Lines changed: 12 additions & 7 deletions
Original file line numberDiff line numberDiff line change
@@ -1,25 +1,28 @@
11
// RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100 %s -Wuninitialized
2+
// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=99 -DOMP99 -verify=expected,rev -ferror-limit 100 %s -Wuninitialized
23

34
int a;
4-
#pragma omp requires unified_address allocate(a) // expected-note {{unified_address clause previously used here}} expected-note {{unified_address clause previously used here}} expected-note {{unified_address clause previously used here}} expected-note {{unified_address clause previously used here}} expected-note {{unified_address clause previously used here}} expected-note{{unified_address clause previously used here}} expected-error {{unexpected OpenMP clause 'allocate' in directive '#pragma omp requires'}}
5+
#pragma omp requires unified_address allocate(a) // rev-note {{unified_address clause previously used here}} expected-note {{unified_address clause previously used here}} expected-note {{unified_address clause previously used here}} expected-note {{unified_address clause previously used here}} expected-note {{unified_address clause previously used here}} expected-note{{unified_address clause previously used here}} expected-error {{unexpected OpenMP clause 'allocate' in directive '#pragma omp requires'}}
56

6-
#pragma omp requires unified_shared_memory // expected-note {{unified_shared_memory clause previously used here}} expected-note{{unified_shared_memory clause previously used here}}
7+
#pragma omp requires unified_shared_memory // rev-note {{unified_shared_memory clause previously used here}} expected-note{{unified_shared_memory clause previously used here}}
78

89
#pragma omp requires unified_shared_memory, unified_shared_memory // expected-error {{Only one unified_shared_memory clause can appear on a requires directive in a single translation unit}} expected-error {{directive '#pragma omp requires' cannot contain more than one 'unified_shared_memory' clause}}
910

1011
#pragma omp requires unified_address // expected-error {{Only one unified_address clause can appear on a requires directive in a single translation unit}}
1112

1213
#pragma omp requires unified_address, unified_address // expected-error {{Only one unified_address clause can appear on a requires directive in a single translation unit}} expected-error {{directive '#pragma omp requires' cannot contain more than one 'unified_address' clause}}
1314

14-
#pragma omp requires reverse_offload // expected-note {{reverse_offload clause previously used here}} expected-note {{reverse_offload clause previously used here}}
15+
#ifdef OMP99
16+
#pragma omp requires reverse_offload // rev-note {{reverse_offload clause previously used here}} rev-note {{reverse_offload clause previously used here}}
1517

16-
#pragma omp requires reverse_offload, reverse_offload // expected-error {{Only one reverse_offload clause can appear on a requires directive in a single translation unit}} expected-error {{directive '#pragma omp requires' cannot contain more than one 'reverse_offload' clause}}
18+
#pragma omp requires reverse_offload, reverse_offload // rev-error {{Only one reverse_offload clause can appear on a requires directive in a single translation unit}} rev-error {{directive '#pragma omp requires' cannot contain more than one 'reverse_offload' clause}}
19+
#endif
1720

18-
#pragma omp requires dynamic_allocators // expected-note {{dynamic_allocators clause previously used here}} expected-note {{dynamic_allocators clause previously used here}}
21+
#pragma omp requires dynamic_allocators // rev-note {{dynamic_allocators clause previously used here}} expected-note {{dynamic_allocators clause previously used here}}
1922

2023
#pragma omp requires dynamic_allocators, dynamic_allocators // expected-error {{Only one dynamic_allocators clause can appear on a requires directive in a single translation unit}} expected-error {{directive '#pragma omp requires' cannot contain more than one 'dynamic_allocators' clause}}
2124

22-
#pragma omp requires atomic_default_mem_order(seq_cst) // expected-note {{atomic_default_mem_order clause previously used here}} expected-note {{atomic_default_mem_order clause previously used here}} expected-note {{atomic_default_mem_order clause previously used here}} expected-note {{atomic_default_mem_order clause previously used here}} expected-note {{atomic_default_mem_order clause previously used here}}
25+
#pragma omp requires atomic_default_mem_order(seq_cst) // rev-note {{atomic_default_mem_order clause previously used here}} expected-note {{atomic_default_mem_order clause previously used here}} expected-note {{atomic_default_mem_order clause previously used here}} expected-note {{atomic_default_mem_order clause previously used here}} expected-note {{atomic_default_mem_order clause previously used here}}
2326

2427
#pragma omp requires atomic_default_mem_order(acq_rel) // expected-error {{Only one atomic_default_mem_order clause can appear on a requires directive in a single translation unit}}
2528

@@ -47,7 +50,9 @@ int a;
4750

4851
#pragma omp requires invalid_clause unified_address // expected-warning {{extra tokens at the end of '#pragma omp requires' are ignored}} expected-error {{expected at least one clause on '#pragma omp requires' directive}}
4952

50-
#pragma omp requires unified_shared_memory, unified_address, reverse_offload, dynamic_allocators, atomic_default_mem_order(seq_cst) // expected-error {{Only one unified_shared_memory clause can appear on a requires directive in a single translation unit}} expected-error{{Only one unified_address clause can appear on a requires directive in a single translation unit}} expected-error{{Only one reverse_offload clause can appear on a requires directive in a single translation unit}} expected-error{{Only one dynamic_allocators clause can appear on a requires directive in a single translation unit}} expected-error {{Only one atomic_default_mem_order clause can appear on a requires directive in a single translation unit}}
53+
#ifdef OMP99
54+
#pragma omp requires unified_shared_memory, unified_address, reverse_offload, dynamic_allocators, atomic_default_mem_order(seq_cst) // rev-error {{Only one unified_shared_memory clause can appear on a requires directive in a single translation unit}} rev-error{{Only one unified_address clause can appear on a requires directive in a single translation unit}} rev-error{{Only one reverse_offload clause can appear on a requires directive in a single translation unit}} rev-error{{Only one dynamic_allocators clause can appear on a requires directive in a single translation unit}} rev-error {{Only one atomic_default_mem_order clause can appear on a requires directive in a single translation unit}}
55+
#endif
5156

5257
namespace A {
5358
#pragma omp requires unified_address // expected-error {{Only one unified_address clause can appear on a requires directive in a single translation unit}}

‎clang/test/OpenMP/requires_target_messages.cpp‎

Lines changed: 4 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,8 +1,9 @@
11
// RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100 %s
2+
// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=99 -DOMP99 -verify=expected,rev -ferror-limit 100 %s
23

34
void foo2() {
45
int a;
5-
#pragma omp target // expected-note 4 {{'target' previously encountered here}}
6+
#pragma omp target // expected-note 3 {{'target' previously encountered here}} rev-note {{'target' previously encountered here}}
67
{
78
a = a + 1;
89
}
@@ -11,5 +12,7 @@ void foo2() {
1112
#pragma omp requires atomic_default_mem_order(seq_cst)
1213
#pragma omp requires unified_address //expected-error {{'target' region encountered before requires directive with 'unified_address' clause}}
1314
#pragma omp requires unified_shared_memory //expected-error {{'target' region encountered before requires directive with 'unified_shared_memory' clause}}
15+
#ifdef OMP99
1416
#pragma omp requires reverse_offload //expected-error {{'target' region encountered before requires directive with 'reverse_offload' clause}}
17+
#endif
1518
#pragma omp requires dynamic_allocators //expected-error {{'target' region encountered before requires directive with 'dynamic_allocators' clause}}

‎clang/test/OpenMP/target_ast_print.cpp‎

Lines changed: 19 additions & 6 deletions
Original file line numberDiff line numberDiff line change
@@ -342,7 +342,18 @@ int main (int argc, char **argv) {
342342
// RUN: %clang_cc1 -DOMP5 -verify -fopenmp-simd -fopenmp-version=50 -ast-print %s | FileCheck %s --check-prefix OMP5
343343
// RUN: %clang_cc1 -DOMP5 -fopenmp-simd -fopenmp-version=50 -x c++ -std=c++11 -emit-pch -o %t %s
344344
// RUN: %clang_cc1 -DOMP5 -fopenmp-simd -fopenmp-version=50 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s --check-prefix OMP5
345+
346+
// RUN: %clang_cc1 -DOMP5 -verify -fopenmp -fopenmp-version=99 -DOMP99 -ast-print %s | FileCheck %s --check-prefixes=OMP5,REV
347+
// RUN: %clang_cc1 -DOMP5 -fopenmp -fopenmp-version=99 -DOMP99 -x c++ -std=c++11 -emit-pch -o %t %s
348+
// RUN: %clang_cc1 -DOMP5 -fopenmp -fopenmp-version=99 -DOMP99 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s --check-prefixes=OMP5,REV
349+
350+
// RUN: %clang_cc1 -DOMP5 -verify -fopenmp-simd -fopenmp-version=99 -DOMP99 -ast-print %s | FileCheck %s --check-prefixes=OMP5,REV
351+
// RUN: %clang_cc1 -DOMP5 -fopenmp-simd -fopenmp-version=99 -DOMP99 -x c++ -std=c++11 -emit-pch -o %t %s
352+
// RUN: %clang_cc1 -DOMP5 -fopenmp-simd -fopenmp-version=99 -DOMP99 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s --check-prefixes=OMP5,REV
353+
354+
#ifdef OMP99
345355
#pragma omp requires reverse_offload
356+
#endif
346357
typedef void **omp_allocator_handle_t;
347358
extern const omp_allocator_handle_t omp_null_allocator;
348359
extern const omp_allocator_handle_t omp_default_mem_alloc;
@@ -370,8 +381,10 @@ T tmain(T argc, T *argv) {
370381
foo();
371382
#pragma omp target if (target:argc > 0) device(device_num: C)
372383
foo();
384+
#ifdef OMP99
373385
#pragma omp target if (C) device(ancestor: argc)
374386
foo();
387+
#endif
375388
#pragma omp target map(i)
376389
foo();
377390
#pragma omp target map(a[0:10], i)
@@ -475,8 +488,8 @@ T tmain(T argc, T *argv) {
475488
// OMP5-NEXT: foo();
476489
// OMP5-NEXT: #pragma omp target if(target: argc > 0) device(device_num: C)
477490
// OMP5-NEXT: foo()
478-
// OMP5-NEXT: #pragma omp target if(C) device(ancestor: argc)
479-
// OMP5-NEXT: foo()
491+
// REV: #pragma omp target if(C) device(ancestor: argc)
492+
// REV-NEXT: foo()
480493
// OMP5-NEXT: #pragma omp target map(tofrom: i)
481494
// OMP5-NEXT: foo()
482495
// OMP5-NEXT: #pragma omp target map(tofrom: a[0:10],i)
@@ -571,8 +584,8 @@ T tmain(T argc, T *argv) {
571584
// OMP5-NEXT: foo();
572585
// OMP5-NEXT: #pragma omp target if(target: argc > 0)
573586
// OMP5-NEXT: foo()
574-
// OMP5-NEXT: #pragma omp target if(5)
575-
// OMP5-NEXT: foo()
587+
// REV: #pragma omp target if(5)
588+
// REV-NEXT: foo()
576589
// OMP5-NEXT: #pragma omp target map(tofrom: i)
577590
// OMP5-NEXT: foo()
578591
// OMP5-NEXT: #pragma omp target map(tofrom: a[0:10],i)
@@ -667,8 +680,8 @@ T tmain(T argc, T *argv) {
667680
// OMP5-NEXT: foo();
668681
// OMP5-NEXT: #pragma omp target if(target: argc > 0) device(device_num: 1)
669682
// OMP5-NEXT: foo()
670-
// OMP5-NEXT: #pragma omp target if(1) device(ancestor: argc)
671-
// OMP5-NEXT: foo()
683+
// REV: #pragma omp target if(1) device(ancestor: argc)
684+
// REV-NEXT: foo()
672685
// OMP5-NEXT: #pragma omp target map(tofrom: i)
673686
// OMP5-NEXT: foo()
674687
// OMP5-NEXT: #pragma omp target map(tofrom: a[0:10],i)

‎clang/test/OpenMP/target_device_codegen.cpp‎

Lines changed: 19 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -6,12 +6,25 @@
66
// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
77
// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s
88
// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -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
9+
10+
// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=99 -DOMP99 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefixes=CHECK,REV
11+
// RUN: %clang_cc1 -fopenmp -fopenmp-version=99 -DOMP99 -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s
12+
// RUN: %clang_cc1 -fopenmp -fopenmp-version=99 -DOMP99 -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-prefixes=CHECK,REV
13+
14+
// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=99 -DOMP99 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefixes=SIMD-ONLY0
15+
// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=99 -DOMP99 -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s
16+
// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=99 -DOMP99 -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-prefixes=SIMD-ONLY0
17+
918
// SIMD-ONLY0-NOT: {{__kmpc|__tgt}}
1019

1120
// expected-no-diagnostics
1221
#ifndef HEADER
1322
#define HEADER
23+
24+
#ifdef OMP99
1425
#pragma omp requires reverse_offload
26+
#endif
27+
1528
void foo(int n) {
1629

1730
// CHECK: [[N:%.+]] = load i32, i32* [[N_ADDR:%.+]],
@@ -40,11 +53,14 @@ void foo(int n) {
4053
// CHECK: [[END]]
4154
#pragma omp target device(device_num: n)
4255
;
43-
// CHECK-NOT: call i32 @__tgt_target_mapper(%struct.ident_t* @{{.+}},
44-
// CHECK: call void @__omp_offloading_{{.+}}_l46()
45-
// CHECK-NOT: call i32 @__tgt_target_mapper(%struct.ident_t* @{{.+}},
56+
57+
#ifdef OMP99
58+
// REV-NOT: call i32 @__tgt_target_mapper(%struct.ident_t* @{{.+}},
59+
// REV: call void @__omp_offloading_{{.+}}_l61()
60+
// REV-NOT: call i32 @__tgt_target_mapper(%struct.ident_t* @{{.+}},
4661
#pragma omp target device(ancestor: n)
4762
;
63+
#endif
4864
}
4965

5066
#endif

‎llvm/include/llvm/Frontend/OpenMP/OMP.td‎

Lines changed: 9 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -590,7 +590,15 @@ def OMP_Requires : Directive<"requires"> {
590590
let allowedClauses = [
591591
VersionedClause<OMPC_UnifiedAddress>,
592592
VersionedClause<OMPC_UnifiedSharedMemory>,
593-
VersionedClause<OMPC_ReverseOffload>,
593+
// OpenMP 5.2 Spec: If an implementation is not supporting a requirement
594+
// (reverse offload in this case) then it should give compile-time error
595+
// termination.
596+
// Seeting supported version for reverse_offload to a distant future version
597+
// 9.9 so that its partial support can be tested in the meantime.
598+
//
599+
// TODO: Correct this supprted version number whenever complete
600+
// implementation of reverse_offload is available.
601+
VersionedClause<OMPC_ReverseOffload, 99>,
594602
VersionedClause<OMPC_DynamicAllocators>,
595603
VersionedClause<OMPC_AtomicDefaultMemOrder>
596604
];

0 commit comments

Comments
 (0)