[OpenMP][Clang] Move partial support of reverse offload to a future version

[saiislam]

Summary

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.

[ABataev]

LG

clang/test/OpenMP/requires_ast_print.cpp

 // RUN: %clang_cc1 -verify -fopenmp -ast-print %s | FileCheck %s
 // RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -emit-pch -o %t %s
 // RUN: %clang_cc1 -fopenmp -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s
 
 // RUN: %clang_cc1 -verify -fopenmp-simd -ast-print %s | FileCheck %s
 // RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -emit-pch -o %t %s
 // RUN: %clang_cc1 -fopenmp-simd -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s
+
+// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=99 -DOMP99 -ast-print %s | FileCheck --check-prefixes=CHECK,REV %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=99 -DOMP99 -x c++ -std=c++11 -emit-pch -o %t %s
+// 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
+
+// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=99 -DOMP99 -ast-print %s | FileCheck --check-prefixes=CHECK,REV %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=99 -DOMP99 -x c++ -std=c++11 -emit-pch -o %t %s
+// 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
 // expected-no-diagnostics
 
 #ifndef HEADER
 #define HEADER
 
 #pragma omp requires unified_address 
 // CHECK:#pragma omp requires unified_address
 
 #pragma omp requires unified_shared_memory
 // CHECK:#pragma omp requires unified_shared_memory
 
+#ifdef OMP99
 #pragma omp requires reverse_offload
-// CHECK:#pragma omp requires reverse_offload
+// REV:#pragma omp requires reverse_offload
+#endif
 
 #pragma omp requires dynamic_allocators
 // CHECK:#pragma omp requires dynamic_allocators
 
 #pragma omp requires atomic_default_mem_order(seq_cst)
 // CHECK:#pragma omp requires atomic_default_mem_order(seq_cst)
 
 #endif

clang/test/OpenMP/requires_messages.cpp

 // RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100  %s -Wuninitialized
+// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=99 -DOMP99 -verify=expected,rev -ferror-limit 100  %s -Wuninitialized
 
 int a;
-#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'}}
+#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'}}
 
-#pragma omp requires unified_shared_memory // expected-note {{unified_shared_memory clause previously used here}} expected-note{{unified_shared_memory clause previously used here}}
+#pragma omp requires unified_shared_memory // rev-note {{unified_shared_memory clause previously used here}} expected-note{{unified_shared_memory clause previously used here}}
 
 #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}}
 
 #pragma omp requires unified_address // expected-error {{Only one unified_address clause can appear on a requires directive in a single translation unit}}
 
 #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}}
 
-#pragma omp requires reverse_offload // expected-note {{reverse_offload clause previously used here}} expected-note {{reverse_offload clause previously used here}}
+#ifdef OMP99
+#pragma omp requires reverse_offload // rev-note {{reverse_offload clause previously used here}} rev-note {{reverse_offload clause previously used here}}
 
-#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}}
+#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}}
+#endif
 
-#pragma omp requires dynamic_allocators // expected-note {{dynamic_allocators clause previously used here}} expected-note {{dynamic_allocators clause previously used here}}
+#pragma omp requires dynamic_allocators // rev-note {{dynamic_allocators clause previously used here}} expected-note {{dynamic_allocators clause previously used here}}
 
 #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}}
 
-#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}}
+#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}}
 
 #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}}
 
 #pragma omp requires atomic_default_mem_order(relaxed) // expected-error {{Only one atomic_default_mem_order clause can appear on a requires directive in a single translation unit}}
 
 #pragma omp requires atomic_default_mem_order // expected-error {{expected '(' after 'atomic_default_mem_order'}} expected-error {{expected at least one clause on '#pragma omp requires' directive}}
 
 #pragma omp requires atomic_default_mem_order( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected 'seq_cst', 'acq_rel' or 'relaxed' in OpenMP clause 'atomic_default_mem_order'}} expected-error {{expected at least one clause on '#pragma omp requires' directive}}

 #pragma omp requires invalid_clause // 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}}
 
 #pragma omp requires nowait // expected-error {{unexpected OpenMP clause 'nowait' in directive '#pragma omp requires'}} expected-error {{expected at least one clause on '#pragma omp requires' directive}}
 
 #pragma omp requires unified_address, invalid_clause // expected-warning {{extra tokens at the end of '#pragma omp requires' are ignored}} expected-error {{Only one unified_address clause can appear on a requires directive in a single translation unit}}
 
 #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}}
 
-#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}}
+#ifdef OMP99
+#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}}
+#endif
 
 namespace A {
   #pragma omp requires unified_address // expected-error {{Only one unified_address clause can appear on a requires directive in a single translation unit}}
   namespace B {
     #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

 // RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100  %s
+// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=99 -DOMP99 -verify=expected,rev -ferror-limit 100  %s
 
 void foo2() {
   int a;
-  #pragma omp target // expected-note 4 {{'target' previously encountered here}}
+  #pragma omp target // expected-note 3 {{'target' previously encountered here}} rev-note {{'target' previously encountered here}}
   {
     a = a + 1;
   }
 }
 
 #pragma omp requires atomic_default_mem_order(seq_cst)
 #pragma omp requires unified_address //expected-error {{'target' region encountered before requires directive with 'unified_address' clause}}
 #pragma omp requires unified_shared_memory //expected-error {{'target' region encountered before requires directive with 'unified_shared_memory' clause}}
+#ifdef OMP99
 #pragma omp requires reverse_offload //expected-error {{'target' region encountered before requires directive with 'reverse_offload' clause}}
+#endif
 #pragma omp requires dynamic_allocators //expected-error {{'target' region encountered before requires directive with 'dynamic_allocators' clause}}

clang/test/OpenMP/target_ast_print.cpp

 ///==========================================================================///
 // RUN: %clang_cc1 -DOMP5 -verify -fopenmp -fopenmp-version=50 -ast-print %s | FileCheck %s --check-prefix OMP5
 // RUN: %clang_cc1 -DOMP5 -fopenmp -fopenmp-version=50 -x c++ -std=c++11 -emit-pch -o %t %s
 // RUN: %clang_cc1 -DOMP5 -fopenmp -fopenmp-version=50 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s --check-prefix OMP5
 
 // RUN: %clang_cc1 -DOMP5 -verify -fopenmp-simd -fopenmp-version=50 -ast-print %s | FileCheck %s --check-prefix OMP5
 // RUN: %clang_cc1 -DOMP5 -fopenmp-simd -fopenmp-version=50 -x c++ -std=c++11 -emit-pch -o %t %s
 // 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
+
+// RUN: %clang_cc1 -DOMP5 -verify -fopenmp -fopenmp-version=99 -DOMP99 -ast-print %s | FileCheck %s --check-prefixes=OMP5,REV
+// RUN: %clang_cc1 -DOMP5 -fopenmp -fopenmp-version=99 -DOMP99 -x c++ -std=c++11 -emit-pch -o %t %s
+// 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
+
+// RUN: %clang_cc1 -DOMP5 -verify -fopenmp-simd -fopenmp-version=99 -DOMP99 -ast-print %s | FileCheck %s --check-prefixes=OMP5,REV
+// RUN: %clang_cc1 -DOMP5 -fopenmp-simd -fopenmp-version=99 -DOMP99 -x c++ -std=c++11 -emit-pch -o %t %s
+// 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
+
+#ifdef OMP99
 #pragma omp requires reverse_offload
+#endif
 typedef void **omp_allocator_handle_t;
 extern const omp_allocator_handle_t omp_null_allocator;
 extern const omp_allocator_handle_t omp_default_mem_alloc;
 extern const omp_allocator_handle_t omp_large_cap_mem_alloc;
 extern const omp_allocator_handle_t omp_const_mem_alloc;
 extern const omp_allocator_handle_t omp_high_bw_mem_alloc;
 extern const omp_allocator_handle_t omp_low_lat_mem_alloc;
 extern const omp_allocator_handle_t omp_cgroup_mem_alloc;

 
 template <typename T, int C>
 T tmain(T argc, T *argv) {
   T i, j, a[20], always, close;
 #pragma omp target device(argc)
   foo();
 #pragma omp target if (target:argc > 0) device(device_num: C)
   foo();
+#ifdef OMP99
 #pragma omp target if (C) device(ancestor: argc)
   foo();
+#endif
 #pragma omp target map(i)
   foo();
 #pragma omp target map(a[0:10], i)
   foo();
 #pragma omp target map(to: i) map(from: j)
   foo();
 #pragma omp target map(always,alloc: i)
   foo();

 }
 
 // OMP5: template <typename T, int C> T tmain(T argc, T *argv) {
 // OMP5-NEXT: T i, j, a[20]
 // OMP5-NEXT: #pragma omp target device(argc){{$}}
 // OMP5-NEXT: foo();
 // OMP5-NEXT: #pragma omp target if(target: argc > 0) device(device_num: C)
 // OMP5-NEXT: foo()
-// OMP5-NEXT: #pragma omp target if(C) device(ancestor: argc)
-// OMP5-NEXT: foo()
+// REV: #pragma omp target if(C) device(ancestor: argc)
+// REV-NEXT: foo()
 // OMP5-NEXT: #pragma omp target map(tofrom: i)
 // OMP5-NEXT: foo()
 // OMP5-NEXT: #pragma omp target map(tofrom: a[0:10],i)
 // OMP5-NEXT: foo()
 // OMP5-NEXT: #pragma omp target map(to: i) map(from: j)
 // OMP5-NEXT: foo()
 // OMP5-NEXT: #pragma omp target map(always,alloc: i)
 // OMP5-NEXT: foo()

 // OMP5-NEXT: #pragma omp target defaultmap(tofrom: aggregate) defaultmap(to: pointer) defaultmap(alloc: scalar)
 // OMP5-NEXT: foo()
 // OMP5: template<> int tmain<int, 5>(int argc, int *argv) {
 // OMP5-NEXT: int i, j, a[20]
 // OMP5-NEXT: #pragma omp target
 // OMP5-NEXT: foo();
 // OMP5-NEXT: #pragma omp target if(target: argc > 0)
 // OMP5-NEXT: foo()
-// OMP5-NEXT: #pragma omp target if(5)
-// OMP5-NEXT: foo()
+// REV: #pragma omp target if(5)
+// REV-NEXT: foo()
 // OMP5-NEXT: #pragma omp target map(tofrom: i)
 // OMP5-NEXT: foo()
 // OMP5-NEXT: #pragma omp target map(tofrom: a[0:10],i)
 // OMP5-NEXT: foo()
 // OMP5-NEXT: #pragma omp target map(to: i) map(from: j)
 // OMP5-NEXT: foo()
 // OMP5-NEXT: #pragma omp target map(always,alloc: i)
 // OMP5-NEXT: foo()

 // OMP5-NEXT: #pragma omp target defaultmap(tofrom: aggregate) defaultmap(to: pointer) defaultmap(alloc: scalar)
 // OMP5-NEXT: foo()
 // OMP5: template<> char tmain<char, 1>(char argc, char *argv) {
 // OMP5-NEXT: char i, j, a[20]
 // OMP5-NEXT: #pragma omp target device(argc)
 // OMP5-NEXT: foo();
 // OMP5-NEXT: #pragma omp target if(target: argc > 0) device(device_num: 1)
 // OMP5-NEXT: foo()
-// OMP5-NEXT: #pragma omp target if(1) device(ancestor: argc)
-// OMP5-NEXT: foo()
+// REV: #pragma omp target if(1) device(ancestor: argc)
+// REV-NEXT: foo()
 // OMP5-NEXT: #pragma omp target map(tofrom: i)
 // OMP5-NEXT: foo()
 // OMP5-NEXT: #pragma omp target map(tofrom: a[0:10],i)
 // OMP5-NEXT: foo()
 // OMP5-NEXT: #pragma omp target map(to: i) map(from: j)
 // OMP5-NEXT: foo()
 // OMP5-NEXT: #pragma omp target map(always,alloc: i)
 // OMP5-NEXT: foo()

clang/test/OpenMP/target_device_codegen.cpp

 // Test host codegen.
 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck %s
 // RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s
 // RUN: %clang_cc1 -fopenmp -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 %s
 
 // 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
 // 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
 // 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
+
+// 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
+// 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
+// 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
+
+// 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
+// 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
+// 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
+
 // SIMD-ONLY0-NOT: {{__kmpc|__tgt}}
 
 // expected-no-diagnostics
 #ifndef HEADER
 #define HEADER
+
+#ifdef OMP99
 #pragma omp requires reverse_offload
+#endif
+
 void foo(int n) {
 
   // CHECK:       [[N:%.+]] = load i32, i32* [[N_ADDR:%.+]],
   // CHECK:       store i32 [[N]], i32* [[DEVICE_CAP:%.+]],
   // CHECK:       [[DEV:%.+]] = load i32, i32* [[DEVICE_CAP]],
   // CHECK:       [[DEVICE:%.+]] = sext i32 [[DEV]] to i64
   // CHECK:       [[RET:%.+]] = call i32 @__tgt_target_mapper(%struct.ident_t* @{{.+}}, i64 [[DEVICE]], i8* @{{[^,]+}}, i32 0, i8** null, i8** null, i64* null, i64* null, i8** null, i8** null)
   // CHECK-NEXT:  [[ERROR:%.+]] = icmp ne i32 [[RET]], 0
void foo(int n) {
   // CHECK-NEXT:  [[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]]
   #pragma omp target device(device_num: n)
   ;
-  // CHECK-NOT:   call i32 @__tgt_target_mapper(%struct.ident_t* @{{.+}},
-  // CHECK:       call void @__omp_offloading_{{.+}}_l46()
-  // CHECK-NOT:   call i32 @__tgt_target_mapper(%struct.ident_t* @{{.+}},
+
+#ifdef OMP99
+  // REV-NOT:   call i32 @__tgt_target_mapper(%struct.ident_t* @{{.+}},
+  // REV:       call void @__omp_offloading_{{.+}}_l61()
+  // REV-NOT:   call i32 @__tgt_target_mapper(%struct.ident_t* @{{.+}},
   #pragma omp target device(ancestor: n)
   ;
+#endif
 }
 
 #endif

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

def OMP_Cancel : Directive<"cancel"> {
   let allowedClauses = [
     VersionedClause<OMPC_If>
   ];
 }
 def OMP_Requires : Directive<"requires"> {
   let allowedClauses = [
     VersionedClause<OMPC_UnifiedAddress>,
     VersionedClause<OMPC_UnifiedSharedMemory>,
-    VersionedClause<OMPC_ReverseOffload>,
+    // OpenMP 5.2 Spec: If an implementation is not supporting a requirement
+    // (reverse offload in this case) then it should give compile-time error
+    // termination.
+    // Seeting supported version for reverse_offload to a distant future version
+    // 9.9 so that its partial support can be tested in the meantime.
+    //
+    // TODO: Correct this supprted version number whenever complete
+    // implementation of reverse_offload is available.
+    VersionedClause<OMPC_ReverseOffload, 99>,
     VersionedClause<OMPC_DynamicAllocators>,
     VersionedClause<OMPC_AtomicDefaultMemOrder>
   ];
 }
 def OMP_TargetData : Directive<"target data"> {
   let allowedClauses = [
     VersionedClause<OMPC_UseDevicePtr>,
     VersionedClause<OMPC_UseDeviceAddr, 50>