Skip to content

Commit 490902e

Browse files
erichkeaneerichkeane
committed
FIX modifier emission for 'create'
1 parent f8a3532 commit 490902e

File tree

2 files changed

+11
-10
lines changed

2 files changed

+11
-10
lines changed

‎clang/lib/CIR/CodeGen/CIRGenDeclOpenACC.cpp‎

Lines changed: 2 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -250,7 +250,8 @@ class OpenACCGlobalDeclareClauseEmitter final
250250
void VisitCreateClause(const OpenACCCreateClause &clause) {
251251
for (const Expr *var : clause.getVarList())
252252
cgm.emitGlobalOpenACCDeclareDataOperands<mlir::acc::CreateOp>(
253-
var, mlir::acc::DataClause::acc_create, {}, /*structured=*/true,
253+
var, mlir::acc::DataClause::acc_create, clause.getModifierList(),
254+
/*structured=*/true,
254255
/*implicit=*/false, /*requiresDtor=*/true);
255256
}
256257
};

‎clang/test/CIR/CodeGenOpenACC/declare-create.cpp‎

Lines changed: 9 additions & 9 deletions
Original file line numberDiff line numberDiff line change
@@ -75,29 +75,29 @@ int NSInt1;
7575
#pragma acc declare create(zero: NSHSE1, NSInt1, NSHSEArr[1:1])
7676
// CHECK: acc.global_ctor @{{.*}}NSHSE1{{.*}}_acc_ctor {
7777
// CHECK-NEXT: %[[GET_GLOBAL:.*]] = cir.get_global @{{.*}}NSHSE1{{.*}} : !cir.ptr<!rec_HasSideEffects>
78-
// CHECK-NEXT: %[[CREATE:.*]] = acc.create varPtr(%[[GET_GLOBAL]] : !cir.ptr<!rec_HasSideEffects>) -> !cir.ptr<!rec_HasSideEffects> {name = "NSHSE1"}
78+
// CHECK-NEXT: %[[CREATE:.*]] = acc.create varPtr(%[[GET_GLOBAL]] : !cir.ptr<!rec_HasSideEffects>) -> !cir.ptr<!rec_HasSideEffects> {modifiers = #acc<data_clause_modifier zero>, name = "NSHSE1"}
7979
// CHECK-NEXT: acc.declare_enter dataOperands(%[[CREATE]] : !cir.ptr<!rec_HasSideEffects>)
8080
// CHECK-NEXT: acc.terminator
8181
// CHECK-NEXT: }
8282
// CHECK: acc.global_dtor @{{.*}}NSHSE1{{.*}}_acc_dtor {
8383
// CHECK-NEXT: %[[GET_GLOBAL:.*]] = cir.get_global @{{.*}}NSHSE1{{.*}} : !cir.ptr<!rec_HasSideEffects>
84-
// CHECK-NEXT: %[[GDP:.*]] = acc.getdeviceptr varPtr(%[[GET_GLOBAL]] : !cir.ptr<!rec_HasSideEffects>) -> !cir.ptr<!rec_HasSideEffects> {dataClause = #acc<data_clause acc_create>, name = "NSHSE1"}
84+
// CHECK-NEXT: %[[GDP:.*]] = acc.getdeviceptr varPtr(%[[GET_GLOBAL]] : !cir.ptr<!rec_HasSideEffects>) -> !cir.ptr<!rec_HasSideEffects> {dataClause = #acc<data_clause acc_create>, modifiers = #acc<data_clause_modifier zero>, name = "NSHSE1"}
8585
// CHECK-NEXT: acc.declare_exit dataOperands(%[[GDP]] : !cir.ptr<!rec_HasSideEffects>)
86-
// CHECK-NEXT: acc.delete accPtr(%[[GDP]] : !cir.ptr<!rec_HasSideEffects>) {dataClause = #acc<data_clause acc_create>, name = "NSHSE1"}
86+
// CHECK-NEXT: acc.delete accPtr(%[[GDP]] : !cir.ptr<!rec_HasSideEffects>) {dataClause = #acc<data_clause acc_create>, modifiers = #acc<data_clause_modifier zero>, name = "NSHSE1"}
8787
// CHECK-NEXT: acc.terminator
8888
// CHECK-NEXT: }
8989
//
9090
// CHECK: acc.global_ctor @{{.*}}NSInt1{{.*}}_acc_ctor {
9191
// CHECK-NEXT: %[[GET_GLOBAL:.*]] = cir.get_global @{{.*}}NSInt1{{.*}} : !cir.ptr<!s32i>
92-
// CHECK-NEXT: %[[CREATE:.*]] = acc.create varPtr(%[[GET_GLOBAL]] : !cir.ptr<!s32i>) -> !cir.ptr<!s32i> {name = "NSInt1"}
92+
// CHECK-NEXT: %[[CREATE:.*]] = acc.create varPtr(%[[GET_GLOBAL]] : !cir.ptr<!s32i>) -> !cir.ptr<!s32i> {modifiers = #acc<data_clause_modifier zero>, name = "NSInt1"}
9393
// CHECK-NEXT: acc.declare_enter dataOperands(%[[CREATE]] : !cir.ptr<!s32i>)
9494
// CHECK-NEXT: acc.terminator
9595
// CHECK-NEXT: }
9696
// CHECK: acc.global_dtor @{{.*}}NSInt1{{.*}}_acc_dtor {
9797
// CHECK-NEXT: %[[GET_GLOBAL:.*]] = cir.get_global @{{.*}}NSInt1{{.*}} : !cir.ptr<!s32i>
98-
// CHECK-NEXT: %[[GDP:.*]] = acc.getdeviceptr varPtr(%[[GET_GLOBAL]] : !cir.ptr<!s32i>) -> !cir.ptr<!s32i> {dataClause = #acc<data_clause acc_create>, name = "NSInt1"}
98+
// CHECK-NEXT: %[[GDP:.*]] = acc.getdeviceptr varPtr(%[[GET_GLOBAL]] : !cir.ptr<!s32i>) -> !cir.ptr<!s32i> {dataClause = #acc<data_clause acc_create>, modifiers = #acc<data_clause_modifier zero>, name = "NSInt1"}
9999
// CHECK-NEXT: acc.declare_exit dataOperands(%[[GDP]] : !cir.ptr<!s32i>)
100-
// CHECK-NEXT: acc.delete accPtr(%[[GDP]] : !cir.ptr<!s32i>) {dataClause = #acc<data_clause acc_create>, name = "NSInt1"}
100+
// CHECK-NEXT: acc.delete accPtr(%[[GDP]] : !cir.ptr<!s32i>) {dataClause = #acc<data_clause acc_create>, modifiers = #acc<data_clause_modifier zero>, name = "NSInt1"}
101101
// CHECK-NEXT: acc.terminator
102102
// CHECK-NEXT: }
103103
//
@@ -110,7 +110,7 @@ int NSInt1;
110110
// CHECK-NEXT: %[[STRIDE:.*]] = arith.constant 1 : i64
111111
// CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[LB_CAST]] : si32) extent(%[[UB_CAST]] : si32) stride(%[[STRIDE]] : i64) startIdx(%[[IDX]] : i64)
112112
// CHECK-NEXT: %[[GET_GLOBAL:.*]] = cir.get_global @{{.*}}NSHSEArr{{.*}} : !cir.ptr<!cir.array<!rec_HasSideEffects x 5>>
113-
// CHECK-NEXT: %[[CREATE:.*]] = acc.create varPtr(%[[GET_GLOBAL]] : !cir.ptr<!cir.array<!rec_HasSideEffects x 5>>) bounds(%[[BOUNDS]]) -> !cir.ptr<!cir.array<!rec_HasSideEffects x 5>> {name = "NSHSEArr[1:1]"}
113+
// CHECK-NEXT: %[[CREATE:.*]] = acc.create varPtr(%[[GET_GLOBAL]] : !cir.ptr<!cir.array<!rec_HasSideEffects x 5>>) bounds(%[[BOUNDS]]) -> !cir.ptr<!cir.array<!rec_HasSideEffects x 5>> {modifiers = #acc<data_clause_modifier zero>, name = "NSHSEArr[1:1]"}
114114
// CHECK-NEXT: acc.declare_enter dataOperands(%[[CREATE]] : !cir.ptr<!cir.array<!rec_HasSideEffects x 5>>)
115115
// CHECK-NEXT: acc.terminator
116116
// CHECK-NEXT: }
@@ -123,9 +123,9 @@ int NSInt1;
123123
// CHECK-NEXT: %[[STRIDE:.*]] = arith.constant 1 : i64
124124
// CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[LB_CAST]] : si32) extent(%[[UB_CAST]] : si32) stride(%[[STRIDE]] : i64) startIdx(%[[IDX]] : i64)
125125
// CHECK-NEXT: %[[GET_GLOBAL:.*]] = cir.get_global @{{.*}}NSHSEArr{{.*}} : !cir.ptr<!cir.array<!rec_HasSideEffects x 5>>
126-
// CHECK-NEXT: %[[GDP:.*]] = acc.getdeviceptr varPtr(%[[GET_GLOBAL]] : !cir.ptr<!cir.array<!rec_HasSideEffects x 5>>) bounds(%[[BOUNDS]]) -> !cir.ptr<!cir.array<!rec_HasSideEffects x 5>> {dataClause = #acc<data_clause acc_create>, name = "NSHSEArr[1:1]"}
126+
// CHECK-NEXT: %[[GDP:.*]] = acc.getdeviceptr varPtr(%[[GET_GLOBAL]] : !cir.ptr<!cir.array<!rec_HasSideEffects x 5>>) bounds(%[[BOUNDS]]) -> !cir.ptr<!cir.array<!rec_HasSideEffects x 5>> {dataClause = #acc<data_clause acc_create>, modifiers = #acc<data_clause_modifier zero>, name = "NSHSEArr[1:1]"}
127127
// CHECK-NEXT: acc.declare_exit dataOperands(%[[GDP]] : !cir.ptr<!cir.array<!rec_HasSideEffects x 5>>)
128-
// CHECK-NEXT: acc.delete accPtr(%[[GDP]] : !cir.ptr<!cir.array<!rec_HasSideEffects x 5>>) {dataClause = #acc<data_clause acc_create>, name = "NSHSEArr[1:1]"}
128+
// CHECK-NEXT: acc.delete accPtr(%[[GDP]] : !cir.ptr<!cir.array<!rec_HasSideEffects x 5>>) {dataClause = #acc<data_clause acc_create>, modifiers = #acc<data_clause_modifier zero>, name = "NSHSEArr[1:1]"}
129129
// CHECK-NEXT: acc.terminator
130130
// CHECK-NEXT: }
131131

0 commit comments

Comments
 (0)