// RUN: %clang_cc1 -fopenacc -triple x86_64-linux-gnu -Wno-openacc-self-if-potential-conflict -emit-cir -fclangir -triple x86_64-linux-pc %s -o - | FileCheck %s struct NoCopyConstruct {}; struct CopyConstruct { CopyConstruct() = default; CopyConstruct(const CopyConstruct&); }; struct NonDefaultCtor { NonDefaultCtor(); }; struct HasDtor { ~HasDtor(); }; // CHECK: acc.firstprivate.recipe @firstprivatization__ZTSA5_7HasDtor : !cir.ptr> init { // CHECK-NEXT: ^bb0(%[[ARG:.*]]: !cir.ptr> {{.*}}): // CHECK-NEXT: cir.alloca !cir.array, !cir.ptr>, ["openacc.firstprivate.init"] // CHECK-NEXT: acc.yield // CHECK-NEXT: } copy { // CHECK-NEXT: ^bb0(%[[ARG_FROM:.*]]: !cir.ptr> {{.*}}, %[[ARG_TO:.*]]: !cir.ptr> {{.*}}): // CHECK-NEXT: %[[DECAY_TO:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_TO]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[ZERO:.*]] = cir.const #cir.int<0> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[ZERO]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN7HasDtorC1ERKS_(%[[DECAY_TO]], %[[FROM_OFFSET]]) nothrow : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[ONE]] : !s64i), !cir.ptr // CHECK-NEXT: %[[ONE_2:.*]] = cir.const #cir.int<1> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[ONE_2]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN7HasDtorC1ERKS_(%[[TO_OFFSET]], %[[FROM_OFFSET]]) nothrow : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: %[[TWO:.*]] = cir.const #cir.int<2> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[TWO]] : !s64i), !cir.ptr // CHECK-NEXT: %[[TWO_2:.*]] = cir.const #cir.int<2> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[TWO_2]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN7HasDtorC1ERKS_(%[[TO_OFFSET]], %[[FROM_OFFSET]]) nothrow : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: %[[THREE:.*]] = cir.const #cir.int<3> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[THREE]] : !s64i), !cir.ptr // CHECK-NEXT: %[[THREE_2:.*]] = cir.const #cir.int<3> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[THREE_2]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN7HasDtorC1ERKS_(%[[TO_OFFSET]], %[[FROM_OFFSET]]) nothrow : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: %[[FOUR:.*]] = cir.const #cir.int<4> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[FOUR]] : !s64i), !cir.ptr // CHECK-NEXT: %[[FOUR_2:.*]] = cir.const #cir.int<4> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[FOUR_2]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN7HasDtorC1ERKS_(%[[TO_OFFSET]], %[[FROM_OFFSET]]) nothrow : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: acc.yield // // CHECK-NEXT: } destroy { // CHECK-NEXT: ^bb0(%[[ARG:.*]]: !cir.ptr> {{.*}}): // CHECK-NEXT: %[[LAST_IDX:.*]] = cir.const #cir.int<4> : !u64i // CHECK-NEXT: %[[ARRPTR:.*]] = cir.cast(array_to_ptrdecay, %[[ARG]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[ELEM:.*]] = cir.ptr_stride(%[[ARRPTR]] : !cir.ptr, %[[LAST_IDX]] : !u64i), !cir.ptr // CHECK-NEXT: %[[ITR:.*]] = cir.alloca !cir.ptr, !cir.ptr>, ["__array_idx"] // CHECK-NEXT: cir.store %[[ELEM]], %[[ITR]] : !cir.ptr, !cir.ptr> // CHECK-NEXT: cir.do { // CHECK-NEXT: %[[ELEM_LOAD:.*]] = cir.load %[[ITR]] : !cir.ptr>, !cir.ptr // CHECK-NEXT: cir.call @_ZN7HasDtorD1Ev(%[[ELEM_LOAD]]) nothrow : (!cir.ptr) -> () // CHECK-NEXT: %[[NEG_ONE:.*]] = cir.const #cir.int<-1> : !s64i // CHECK-NEXT: %[[PREVELEM:.*]] = cir.ptr_stride(%[[ELEM_LOAD]] : !cir.ptr, %[[NEG_ONE]] : !s64i), !cir.ptr // CHECK-NEXT: cir.store %[[PREVELEM]], %[[ITR]] : !cir.ptr, !cir.ptr> // CHECK-NEXT: cir.yield // CHECK-NEXT: } while { // CHECK-NEXT: %[[ELEM_LOAD:.*]] = cir.load %[[ITR]] : !cir.ptr>, !cir.ptr // CHECK-NEXT: %[[CMP:.*]] = cir.cmp(ne, %[[ELEM_LOAD]], %[[ARRPTR]]) : !cir.ptr, !cir.bool // CHECK-NEXT: cir.condition(%[[CMP]]) // CHECK-NEXT: } // CHECK-NEXT: acc.yield // CHECK-NEXT: } // // CHECK-NEXT: acc.firstprivate.recipe @firstprivatization__ZTSA5_14NonDefaultCtor : !cir.ptr> init { // CHECK-NEXT: ^bb0(%[[ARG:.*]]: !cir.ptr> {{.*}}): // CHECK-NEXT: %[[ALLOCA:.*]] = cir.alloca !cir.array, !cir.ptr>, ["openacc.firstprivate.init"] // CHECK-NEXT: acc.yield // CHECK-NEXT: } copy { // CHECK-NEXT: ^bb0(%[[ARG_FORM:.*]]: !cir.ptr> {{.*}}, %[[ARG_TO:.*]]: !cir.ptr> {{.*}}): // CHECK-NEXT: %[[TO_DECAY:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_TO]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[ZERO:.*]] = cir.const #cir.int<0> // CHECK-NEXT: %[[FROM_DECAY:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[FROM_DECAY]] : !cir.ptr, %[[ZERO]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN14NonDefaultCtorC1ERKS_(%[[TO_DECAY]], %[[FROM_OFFSET]]) nothrow : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[ONE]] : !s64i), !cir.ptr // CHECK-NEXT: %[[ONE_2:.*]] = cir.const #cir.int<1> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[ONE_2]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN14NonDefaultCtorC1ERKS_(%[[TO_OFFSET]], %[[FROM_OFFSET]]) nothrow : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: %[[TWO:.*]] = cir.const #cir.int<2> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[TWO]] : !s64i), !cir.ptr // CHECK-NEXT: %[[TWO_2:.*]] = cir.const #cir.int<2> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[TWO_2]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN14NonDefaultCtorC1ERKS_(%[[TO_OFFSET]], %[[FROM_OFFSET]]) nothrow : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: %[[THREE:.*]] = cir.const #cir.int<3> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[THREE]] : !s64i), !cir.ptr // CHECK-NEXT: %[[THREE_2:.*]] = cir.const #cir.int<3> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[THREE_2]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN14NonDefaultCtorC1ERKS_(%[[TO_OFFSET]], %[[FROM_OFFSET]]) nothrow : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: %[[FOUR:.*]] = cir.const #cir.int<4> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[FOUR]] : !s64i), !cir.ptr // CHECK-NEXT: %[[FOUR_2:.*]] = cir.const #cir.int<4> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[FOUR_2]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN14NonDefaultCtorC1ERKS_(%[[TO_OFFSET]], %[[FROM_OFFSET]]) nothrow : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: acc.yield // CHECK-NEXT: } // // CHECK-NEXT: acc.firstprivate.recipe @firstprivatization__ZTSA5_13CopyConstruct : !cir.ptr> init { // CHECK-NEXT: ^bb0(%[[ARG:.*]]: !cir.ptr> {{.*}}): // CHECK-NEXT: cir.alloca !cir.array, !cir.ptr>, ["openacc.firstprivate.init"] // CHECK-NEXT: acc.yield // CHECK-NEXT: } copy { // CHECK-NEXT: ^bb0(%[[ARG_FORM:.*]]: !cir.ptr> {{.*}}, %[[ARG_TO:.*]]: !cir.ptr> {{.*}}): // CHECK-NEXT: %[[TO_DECAY:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_TO]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[ZERO:.*]] = cir.const #cir.int<0> // CHECK-NEXT: %[[FROM_DECAY:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[FROM_DECAY]] : !cir.ptr, %[[ZERO]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN13CopyConstructC1ERKS_(%[[TO_DECAY]], %[[FROM_OFFSET]]) : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[ONE]] : !s64i), !cir.ptr // CHECK-NEXT: %[[ONE_2:.*]] = cir.const #cir.int<1> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[ONE_2]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN13CopyConstructC1ERKS_(%[[TO_OFFSET]], %[[FROM_OFFSET]]) : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: %[[TWO:.*]] = cir.const #cir.int<2> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[TWO]] : !s64i), !cir.ptr // CHECK-NEXT: %[[TWO_2:.*]] = cir.const #cir.int<2> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[TWO_2]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN13CopyConstructC1ERKS_(%[[TO_OFFSET]], %[[FROM_OFFSET]]) : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: %[[THREE:.*]] = cir.const #cir.int<3> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[THREE]] : !s64i), !cir.ptr // CHECK-NEXT: %[[THREE_2:.*]] = cir.const #cir.int<3> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[THREE_2]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN13CopyConstructC1ERKS_(%[[TO_OFFSET]], %[[FROM_OFFSET]]) : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: %[[FOUR:.*]] = cir.const #cir.int<4> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[FOUR]] : !s64i), !cir.ptr // CHECK-NEXT: %[[FOUR_2:.*]] = cir.const #cir.int<4> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[FOUR_2]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN13CopyConstructC1ERKS_(%[[TO_OFFSET]], %[[FROM_OFFSET]]) : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: acc.yield // CHECK-NEXT: } // // CHECK-NEXT: acc.firstprivate.recipe @firstprivatization__ZTSA5_15NoCopyConstruct : !cir.ptr> init { // CHECK-NEXT: ^bb0(%[[ARG:.*]]: !cir.ptr> {{.*}}): // CHECK-NEXT: cir.alloca !cir.array, !cir.ptr>, ["openacc.firstprivate.init"] // CHECK-NEXT: acc.yield // CHECK-NEXT: } copy { // CHECK-NEXT: ^bb0(%[[ARG_FORM:.*]]: !cir.ptr> {{.*}}, %[[ARG_TO:.*]]: !cir.ptr> {{.*}}): // CHECK-NEXT: %[[TO_DECAY:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_TO]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[ZERO:.*]] = cir.const #cir.int<0> // CHECK-NEXT: %[[FROM_DECAY:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[FROM_DECAY]] : !cir.ptr, %[[ZERO]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN15NoCopyConstructC1ERKS_(%[[TO_DECAY]], %[[FROM_OFFSET]]) nothrow : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[ONE]] : !s64i), !cir.ptr // CHECK-NEXT: %[[ONE_2:.*]] = cir.const #cir.int<1> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[ONE_2]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN15NoCopyConstructC1ERKS_(%[[TO_OFFSET]], %[[FROM_OFFSET]]) nothrow : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: %[[TWO:.*]] = cir.const #cir.int<2> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[TWO]] : !s64i), !cir.ptr // CHECK-NEXT: %[[TWO_2:.*]] = cir.const #cir.int<2> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[TWO_2]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN15NoCopyConstructC1ERKS_(%[[TO_OFFSET]], %[[FROM_OFFSET]]) nothrow : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: %[[THREE:.*]] = cir.const #cir.int<3> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[THREE]] : !s64i), !cir.ptr // CHECK-NEXT: %[[THREE_2:.*]] = cir.const #cir.int<3> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[THREE_2]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN15NoCopyConstructC1ERKS_(%[[TO_OFFSET]], %[[FROM_OFFSET]]) nothrow : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: %[[FOUR:.*]] = cir.const #cir.int<4> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[FOUR]] : !s64i), !cir.ptr // CHECK-NEXT: %[[FOUR_2:.*]] = cir.const #cir.int<4> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[FOUR_2]] : !u64i), !cir.ptr // CHECK-NEXT: cir.call @_ZN15NoCopyConstructC1ERKS_(%[[TO_OFFSET]], %[[FROM_OFFSET]]) nothrow : (!cir.ptr, !cir.ptr) -> () // // CHECK-NEXT: acc.yield // CHECK-NEXT: } // // CHECK-NEXT: acc.firstprivate.recipe @firstprivatization__ZTSA5_f : !cir.ptr> init { // CHECK-NEXT: ^bb0(%[[ARG:.*]]: !cir.ptr> {{.*}}): // CHECK-NEXT: cir.alloca !cir.array, !cir.ptr>, ["openacc.firstprivate.init"] // CHECK-NEXT: acc.yield // CHECK-NEXT: } copy { // CHECK-NEXT: ^bb0(%[[ARG_FORM:.*]]: !cir.ptr> {{.*}}, %[[ARG_TO:.*]]: !cir.ptr> {{.*}}): // CHECK-NEXT: %[[TO_DECAY:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_TO]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[ZERO:.*]] = cir.const #cir.int<0> // CHECK-NEXT: %[[FROM_DECAY:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[FROM_DECAY]] : !cir.ptr, %[[ZERO]] : !u64i), !cir.ptr // CHECK-NEXT: %[[FROM_LOAD:.*]] = cir.load {{.*}}%[[FROM_OFFSET]] : !cir.ptr, !cir.float // CHECK-NEXT: cir.store {{.*}} %[[FROM_LOAD]], %[[TO_DECAY]] : !cir.float, !cir.ptr // // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[ONE]] : !s64i), !cir.ptr // CHECK-NEXT: %[[ONE_2:.*]] = cir.const #cir.int<1> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[ONE_2]] : !u64i), !cir.ptr // CHECK-NEXT: %[[FROM_LOAD:.*]] = cir.load {{.*}}%[[FROM_OFFSET]] : !cir.ptr, !cir.float // CHECK-NEXT: cir.store {{.*}} %[[FROM_LOAD]], %[[TO_OFFSET]] : !cir.float, !cir.ptr // // CHECK-NEXT: %[[TWO:.*]] = cir.const #cir.int<2> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[TWO]] : !s64i), !cir.ptr // CHECK-NEXT: %[[TWO_2:.*]] = cir.const #cir.int<2> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[TWO_2]] : !u64i), !cir.ptr // CHECK-NEXT: %[[FROM_LOAD:.*]] = cir.load {{.*}}%[[FROM_OFFSET]] : !cir.ptr, !cir.float // CHECK-NEXT: cir.store {{.*}} %[[FROM_LOAD]], %[[TO_OFFSET]] : !cir.float, !cir.ptr // // CHECK-NEXT: %[[THREE:.*]] = cir.const #cir.int<3> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[THREE]] : !s64i), !cir.ptr // CHECK-NEXT: %[[THREE_2:.*]] = cir.const #cir.int<3> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[THREE_2]] : !u64i), !cir.ptr // CHECK-NEXT: %[[FROM_LOAD:.*]] = cir.load {{.*}}%[[FROM_OFFSET]] : !cir.ptr, !cir.float // CHECK-NEXT: cir.store {{.*}} %[[FROM_LOAD]], %[[TO_OFFSET]] : !cir.float, !cir.ptr // // CHECK-NEXT: %[[FOUR:.*]] = cir.const #cir.int<4> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[FOUR]] : !s64i), !cir.ptr // CHECK-NEXT: %[[FOUR_2:.*]] = cir.const #cir.int<4> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[FOUR_2]] : !u64i), !cir.ptr // CHECK-NEXT: %[[FROM_LOAD:.*]] = cir.load {{.*}}%[[FROM_OFFSET]] : !cir.ptr, !cir.float // CHECK-NEXT: cir.store {{.*}} %[[FROM_LOAD]], %[[TO_OFFSET]] : !cir.float, !cir.ptr // CHECK-NEXT: acc.yield // CHECK-NEXT: } // // CHECK-NEXT: acc.firstprivate.recipe @firstprivatization__ZTSA5_i : !cir.ptr> init { // CHECK-NEXT: ^bb0(%[[ARG:.*]]: !cir.ptr> {{.*}}): // CHECK-NEXT: cir.alloca !cir.array, !cir.ptr>, ["openacc.firstprivate.init"] // CHECK-NEXT: acc.yield // CHECK-NEXT: } copy { // CHECK-NEXT: ^bb0(%[[ARG_FORM:.*]]: !cir.ptr> {{.*}}, %[[ARG_TO:.*]]: !cir.ptr> {{.*}}): // CHECK-NEXT: %[[TO_DECAY:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_TO]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[ZERO:.*]] = cir.const #cir.int<0> // CHECK-NEXT: %[[FROM_DECAY:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[FROM_DECAY]] : !cir.ptr, %[[ZERO]] : !u64i), !cir.ptr // CHECK-NEXT: %[[FROM_LOAD:.*]] = cir.load {{.*}}%[[FROM_OFFSET]] : !cir.ptr, !s32i // CHECK-NEXT: cir.store {{.*}} %[[FROM_LOAD]], %[[TO_DECAY]] : !s32i, !cir.ptr // // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[ONE]] : !s64i), !cir.ptr // CHECK-NEXT: %[[ONE_2:.*]] = cir.const #cir.int<1> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[ONE_2]] : !u64i), !cir.ptr // CHECK-NEXT: %[[FROM_LOAD:.*]] = cir.load {{.*}}%[[FROM_OFFSET]] : !cir.ptr, !s32i // CHECK-NEXT: cir.store {{.*}} %[[FROM_LOAD]], %[[TO_OFFSET]] : !s32i, !cir.ptr // // CHECK-NEXT: %[[TWO:.*]] = cir.const #cir.int<2> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[TWO]] : !s64i), !cir.ptr // CHECK-NEXT: %[[TWO_2:.*]] = cir.const #cir.int<2> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[TWO_2]] : !u64i), !cir.ptr // CHECK-NEXT: %[[FROM_LOAD:.*]] = cir.load {{.*}}%[[FROM_OFFSET]] : !cir.ptr, !s32i // CHECK-NEXT: cir.store {{.*}} %[[FROM_LOAD]], %[[TO_OFFSET]] : !s32i, !cir.ptr // // CHECK-NEXT: %[[THREE:.*]] = cir.const #cir.int<3> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[THREE]] : !s64i), !cir.ptr // CHECK-NEXT: %[[THREE_2:.*]] = cir.const #cir.int<3> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[THREE_2]] : !u64i), !cir.ptr // CHECK-NEXT: %[[FROM_LOAD:.*]] = cir.load {{.*}}%[[FROM_OFFSET]] : !cir.ptr, !s32i // CHECK-NEXT: cir.store {{.*}} %[[FROM_LOAD]], %[[TO_OFFSET]] : !s32i, !cir.ptr // // CHECK-NEXT: %[[FOUR:.*]] = cir.const #cir.int<4> // CHECK-NEXT: %[[TO_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_TO]] : !cir.ptr, %[[FOUR]] : !s64i), !cir.ptr // CHECK-NEXT: %[[FOUR_2:.*]] = cir.const #cir.int<4> // CHECK-NEXT: %[[DECAY_FROM:.*]] = cir.cast(array_to_ptrdecay, %[[ARG_FROM]] : !cir.ptr>), !cir.ptr // CHECK-NEXT: %[[FROM_OFFSET:.*]] = cir.ptr_stride(%[[DECAY_FROM]] : !cir.ptr, %[[FOUR_2]] : !u64i), !cir.ptr // CHECK-NEXT: %[[FROM_LOAD:.*]] = cir.load {{.*}}%[[FROM_OFFSET]] : !cir.ptr, !s32i // CHECK-NEXT: cir.store {{.*}} %[[FROM_LOAD]], %[[TO_OFFSET]] : !s32i, !cir.ptr // CHECK-NEXT: acc.yield // CHECK-NEXT: } // // CHECK-NEXT: acc.firstprivate.recipe @firstprivatization__ZTS7HasDtor : !cir.ptr init { // CHECK-NEXT: ^bb0(%[[ARG:.*]]: !cir.ptr {{.*}}): // CHECK-NEXT: cir.alloca !rec_HasDtor, !cir.ptr, ["openacc.firstprivate.init"] // CHECK-NEXT: acc.yield // CHECK-NEXT: } copy { // CHECK-NEXT: ^bb0(%[[ARG_FROM:.*]]: !cir.ptr {{.*}}, %[[ARG_TO:.*]]: !cir.ptr {{.*}}): // CHECK-NEXT: cir.call @_ZN7HasDtorC1ERKS_(%[[ARG_TO]], %[[ARG_FROM]]) nothrow : (!cir.ptr, !cir.ptr) -> () // CHECK-NEXT: acc.yield // CHECK-NEXT: } destroy { // CHECK-NEXT: ^bb0(%[[ARG:.*]]: !cir.ptr {{.*}}): // CHECK-NEXT: cir.call @_ZN7HasDtorD1Ev(%[[ARG]]) nothrow : (!cir.ptr) -> () // CHECK-NEXT: acc.yield // CHECK-NEXT: } // // CHECK-NEXT: acc.firstprivate.recipe @firstprivatization__ZTS14NonDefaultCtor : !cir.ptr init { // CHECK-NEXT: ^bb0(%[[ARG:.*]]: !cir.ptr {{.*}}): // CHECK-NEXT: %[[ALLOCA:.*]] = cir.alloca !rec_NonDefaultCtor, !cir.ptr, ["openacc.firstprivate.init"] // CHECK-NEXT: acc.yield // CHECK-NEXT: } copy { // CHECK-NEXT: ^bb0(%[[ARG_FROM:.*]]: !cir.ptr {{.*}}, %[[ARG_TO:.*]]: !cir.ptr {{.*}}): // CHECK-NEXT: cir.call @_ZN14NonDefaultCtorC1ERKS_(%[[ARG_TO]], %[[ARG_FROM]]) nothrow : (!cir.ptr, !cir.ptr) -> () // CHECK-NEXT: acc.yield // CHECK-NEXT: } // // CHECK-NEXT: acc.firstprivate.recipe @firstprivatization__ZTS13CopyConstruct : !cir.ptr init { // CHECK-NEXT: ^bb0(%[[ARG:.*]]: !cir.ptr {{.*}}): // CHECK-NEXT: cir.alloca !rec_CopyConstruct, !cir.ptr, ["openacc.firstprivate.init"] // CHECK-NEXT: acc.yield // CHECK-NEXT: } copy { // CHECK-NEXT: ^bb0(%[[ARG_FROM:.*]]: !cir.ptr {{.*}}, %[[ARG_TO:.*]]: !cir.ptr {{.*}}): // CHECK-NEXT: cir.call @_ZN13CopyConstructC1ERKS_(%[[ARG_TO]], %[[ARG_FROM]]) : (!cir.ptr, !cir.ptr) -> () // CHECK-NEXT: acc.yield // CHECK-NEXT: } // // CHECK-NEXT: acc.firstprivate.recipe @firstprivatization__ZTS15NoCopyConstruct : !cir.ptr init { // CHECK-NEXT: ^bb0(%[[ARG:.*]]: !cir.ptr {{.*}}): // CHECK-NEXT: cir.alloca !rec_NoCopyConstruct, !cir.ptr, ["openacc.firstprivate.init"] // CHECK-NEXT: acc.yield // CHECK-NEXT: } copy { // CHECK-NEXT: ^bb0(%[[ARG_FROM:.*]]: !cir.ptr {{.*}}, %[[ARG_TO:.*]]: !cir.ptr {{.*}}): // CHECK-NEXT: cir.call @_ZN15NoCopyConstructC1ERKS_(%[[ARG_TO]], %[[ARG_FROM]]) nothrow : (!cir.ptr, !cir.ptr) -> () // CHECK-NEXT: acc.yield // CHECK-NEXT: } // // CHECK-NEXT: acc.firstprivate.recipe @firstprivatization__ZTSf : !cir.ptr init { // CHECK-NEXT: ^bb0(%[[ARG:.*]]: !cir.ptr {{.*}}): // CHECK-NEXT: cir.alloca !cir.float, !cir.ptr, ["openacc.firstprivate.init"] // CHECK-NEXT: acc.yield // CHECK-NEXT: } copy { // CHECK-NEXT: ^bb0(%[[ARG_FROM:.*]]: !cir.ptr {{.*}}, %[[ARG_TO:.*]]: !cir.ptr {{.*}}): // CHECK-NEXT: %[[FROM_LOAD:.*]] = cir.load {{.*}}%[[ARG_FROM]] : !cir.ptr, !cir.float // CHECK-NEXT: cir.store {{.*}} %[[FROM_LOAD]], %[[ARG_TO]] : !cir.float, !cir.ptr // CHECK-NEXT: acc.yield // CHECK-NEXT: } // // CHECK-NEXT: acc.firstprivate.recipe @firstprivatization__ZTSi : !cir.ptr init { // CHECK-NEXT: ^bb0(%[[ARG:.*]]: !cir.ptr {{.*}}): // CHECK-NEXT: cir.alloca !s32i, !cir.ptr, ["openacc.firstprivate.init"] // CHECK-NEXT: acc.yield // CHECK-NEXT: } copy { // CHECK-NEXT: ^bb0(%[[ARG_FROM:.*]]: !cir.ptr {{.*}}, %[[ARG_TO:.*]]: !cir.ptr {{.*}}): // CHECK-NEXT: %[[FROM_LOAD:.*]] = cir.load {{.*}}%[[ARG_FROM]] : !cir.ptr, !s32i // CHECK-NEXT: cir.store {{.*}} %[[FROM_LOAD]], %[[ARG_TO]] : !s32i, !cir.ptr // CHECK-NEXT: acc.yield // CHECK-NEXT: } extern "C" void acc_compute() { // CHECK: cir.func{{.*}} @acc_compute() { int someInt; // CHECK-NEXT: %[[SOMEINT:.*]] = cir.alloca !s32i, !cir.ptr, ["someInt"] float someFloat; // CHECK-NEXT: %[[SOMEFLOAT:.*]] = cir.alloca !cir.float, !cir.ptr, ["someFloat"] NoCopyConstruct noCopy; // CHECK-NEXT: %[[NOCOPY:.*]] = cir.alloca !rec_NoCopyConstruct, !cir.ptr, ["noCopy"] CopyConstruct hasCopy; // CHECK-NEXT: %[[HASCOPY:.*]] = cir.alloca !rec_CopyConstruct, !cir.ptr, ["hasCopy"] NonDefaultCtor notDefCtor; // CHECK-NEXT: %[[NOTDEFCTOR:.*]] = cir.alloca !rec_NonDefaultCtor, !cir.ptr, ["notDefCtor", init] HasDtor dtor; // CHECK-NEXT: %[[DTOR:.*]] = cir.alloca !rec_HasDtor, !cir.ptr, ["dtor"] int someIntArr[5]; // CHECK-NEXT: %[[INTARR:.*]] = cir.alloca !cir.array, !cir.ptr>, ["someIntArr"] float someFloatArr[5]; // CHECK-NEXT: %[[FLOATARR:.*]] = cir.alloca !cir.array, !cir.ptr>, ["someFloatArr"] NoCopyConstruct noCopyArr[5]; // CHECK-NEXT: %[[NOCOPYARR:.*]] = cir.alloca !cir.array, !cir.ptr>, ["noCopyArr"] CopyConstruct hasCopyArr[5]; // CHECK-NEXT: %[[HASCOPYARR:.*]] = cir.alloca !cir.array, !cir.ptr>, ["hasCopyArr"] NonDefaultCtor notDefCtorArr[5]; // CHECK-NEXT: %[[NOTDEFCTORARR:.*]] = cir.alloca !cir.array, !cir.ptr>, ["notDefCtorArr", init] HasDtor dtorArr[5]; // CHECK-NEXT: %[[DTORARR:.*]] = cir.alloca !cir.array, !cir.ptr>, ["dtorArr"] // CHECK-NEXT: cir.call @_ZN14NonDefaultCtorC1Ev(%[[NOTDEFCTOR]]) : (!cir.ptr) -> () #pragma acc parallel firstprivate(someInt) ; // CHECK: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[SOMEINT]] : !cir.ptr) -> !cir.ptr {name = "someInt"} // CHECK-NEXT: acc.parallel firstprivate(@firstprivatization__ZTSi -> %[[PRIVATE]] : !cir.ptr) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc serial firstprivate(someFloat) ; // CHECK-NEXT: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[SOMEFLOAT]] : !cir.ptr) -> !cir.ptr {name = "someFloat"} // CHECK-NEXT: acc.serial firstprivate(@firstprivatization__ZTSf -> %[[PRIVATE]] : !cir.ptr) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc parallel firstprivate(noCopy) ; // CHECK-NEXT: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[NOCOPY]] : !cir.ptr) -> !cir.ptr {name = "noCopy"} // CHECK-NEXT: acc.parallel firstprivate(@firstprivatization__ZTS15NoCopyConstruct -> %[[PRIVATE]] : !cir.ptr // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc serial firstprivate(hasCopy) ; // CHECK-NEXT: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[HASCOPY]] : !cir.ptr) -> !cir.ptr {name = "hasCopy"} // CHECK-NEXT: acc.serial firstprivate(@firstprivatization__ZTS13CopyConstruct -> %[[PRIVATE]] : !cir.ptr) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc serial firstprivate(notDefCtor) ; // CHECK-NEXT: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[NOTDEFCTOR]] : !cir.ptr) -> !cir.ptr {name = "notDefCtor"} // CHECK-NEXT: acc.serial firstprivate(@firstprivatization__ZTS14NonDefaultCtor -> %[[PRIVATE]] : !cir.ptr) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc serial firstprivate(dtor) ; // CHECK-NEXT: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[DTOR]] : !cir.ptr) -> !cir.ptr {name = "dtor"} // CHECK-NEXT: acc.serial firstprivate(@firstprivatization__ZTS7HasDtor -> %[[PRIVATE]] : !cir.ptr) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc parallel firstprivate(someInt, someFloat, noCopy, hasCopy, notDefCtor, dtor) ; // CHECK: %[[PRIVATE1:.*]] = acc.firstprivate varPtr(%[[SOMEINT]] : !cir.ptr) -> !cir.ptr {name = "someInt"} // CHECK-NEXT: %[[PRIVATE2:.*]] = acc.firstprivate varPtr(%[[SOMEFLOAT]] : !cir.ptr) -> !cir.ptr {name = "someFloat"} // CHECK-NEXT: %[[PRIVATE3:.*]] = acc.firstprivate varPtr(%[[NOCOPY]] : !cir.ptr) -> !cir.ptr {name = "noCopy"} // CHECK-NEXT: %[[PRIVATE4:.*]] = acc.firstprivate varPtr(%[[HASCOPY]] : !cir.ptr) -> !cir.ptr {name = "hasCopy"} // CHECK-NEXT: %[[PRIVATE5:.*]] = acc.firstprivate varPtr(%[[NOTDEFCTOR]] : !cir.ptr) -> !cir.ptr {name = "notDefCtor"} // CHECK-NEXT: %[[PRIVATE6:.*]] = acc.firstprivate varPtr(%[[DTOR]] : !cir.ptr) -> !cir.ptr {name = "dtor"} // CHECK-NEXT: acc.parallel firstprivate(@firstprivatization__ZTSi -> %[[PRIVATE1]] : !cir.ptr, // CHECK-SAME: @firstprivatization__ZTSf -> %[[PRIVATE2]] : !cir.ptr, // CHECK-SAME: @firstprivatization__ZTS15NoCopyConstruct -> %[[PRIVATE3]] : !cir.ptr, // CHECK-SAME: @firstprivatization__ZTS13CopyConstruct -> %[[PRIVATE4]] : !cir.ptr, // CHECK-SAME: @firstprivatization__ZTS14NonDefaultCtor -> %[[PRIVATE5]] : !cir.ptr, // CHECK-SAME: @firstprivatization__ZTS7HasDtor -> %[[PRIVATE6]] : !cir.ptr) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc serial firstprivate(someIntArr[1]) ; // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE_CONST:.*]] = arith.constant 1 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CONST]] : i64) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[INTARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "someIntArr[1]"} // CHECK-NEXT: acc.serial firstprivate(@firstprivatization__ZTSA5_i -> %[[PRIVATE]] : !cir.ptr>) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc parallel firstprivate(someFloatArr[1]) ; // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE_CONST:.*]] = arith.constant 1 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CONST]] : i64) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[FLOATARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "someFloatArr[1]"} // CHECK-NEXT: acc.parallel firstprivate(@firstprivatization__ZTSA5_f -> %[[PRIVATE]] : !cir.ptr>) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc serial firstprivate(noCopyArr[1]) ; // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE_CONST:.*]] = arith.constant 1 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CONST]] : i64) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[NOCOPYARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "noCopyArr[1]"} // CHECK-NEXT: acc.serial firstprivate(@firstprivatization__ZTSA5_15NoCopyConstruct -> %[[PRIVATE]] : !cir.ptr>) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc parallel firstprivate(hasCopyArr[1]) ; // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE_CONST:.*]] = arith.constant 1 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CONST]] : i64) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[HASCOPYARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "hasCopyArr[1]"} // CHECK-NEXT: acc.parallel firstprivate(@firstprivatization__ZTSA5_13CopyConstruct -> %[[PRIVATE]] : !cir.ptr>) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc parallel firstprivate(notDefCtorArr[1]) ; // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE_CONST:.*]] = arith.constant 1 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CONST]] : i64) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[NOTDEFCTORARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "notDefCtorArr[1]"} // CHECK-NEXT: acc.parallel firstprivate(@firstprivatization__ZTSA5_14NonDefaultCtor -> %[[PRIVATE]] : !cir.ptr>) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc parallel firstprivate(dtorArr[1]) ; // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE_CONST:.*]] = arith.constant 1 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CONST]] : i64) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[DTORARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "dtorArr[1]"} // CHECK-NEXT: acc.parallel firstprivate(@firstprivatization__ZTSA5_7HasDtor -> %[[PRIVATE]] : !cir.ptr>) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc serial firstprivate(someIntArr[1], someFloatArr[1], noCopyArr[1], hasCopyArr[1], notDefCtorArr[1], dtorArr[1]) ; // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE_CONST:.*]] = arith.constant 1 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CONST]] : i64) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE1:.*]] = acc.firstprivate varPtr(%[[INTARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "someIntArr[1]"} // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE_CONST:.*]] = arith.constant 1 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CONST]] : i64) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE2:.*]] = acc.firstprivate varPtr(%[[FLOATARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "someFloatArr[1]"} // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE_CONST:.*]] = arith.constant 1 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CONST]] : i64) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE3:.*]] = acc.firstprivate varPtr(%[[NOCOPYARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "noCopyArr[1]"} // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE_CONST:.*]] = arith.constant 1 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CONST]] : i64) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE4:.*]] = acc.firstprivate varPtr(%[[HASCOPYARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "hasCopyArr[1]"} // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE_CONST:.*]] = arith.constant 1 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CONST]] : i64) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE5:.*]] = acc.firstprivate varPtr(%[[NOTDEFCTORARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "notDefCtorArr[1]"} // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE_CONST:.*]] = arith.constant 1 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CONST]] : i64) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE6:.*]] = acc.firstprivate varPtr(%[[DTORARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "dtorArr[1]"} // CHECK-NEXT: acc.serial firstprivate(@firstprivatization__ZTSA5_i -> %[[PRIVATE1]] : !cir.ptr>, // CHECK-SAME: @firstprivatization__ZTSA5_f -> %[[PRIVATE2]] : !cir.ptr>, // CHECK-SAME: @firstprivatization__ZTSA5_15NoCopyConstruct -> %[[PRIVATE3]] : !cir.ptr>, // CHECK-SAME: @firstprivatization__ZTSA5_13CopyConstruct -> %[[PRIVATE4]] : !cir.ptr>, // CHECK-SAME: @firstprivatization__ZTSA5_14NonDefaultCtor -> %[[PRIVATE5]] : !cir.ptr>, // CHECK-SAME: @firstprivatization__ZTSA5_7HasDtor -> %[[PRIVATE6]] : !cir.ptr>) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc parallel firstprivate(someIntArr[1:1]) ; // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST2:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CAST2]] : si32) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[INTARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "someIntArr[1:1]"} // CHECK-NEXT: acc.parallel firstprivate(@firstprivatization__ZTSA5_i -> %[[PRIVATE]] : !cir.ptr>) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc serial firstprivate(someFloatArr[1:1]) ; // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST2:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CAST2]] : si32) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[FLOATARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "someFloatArr[1:1]"} // CHECK-NEXT: acc.serial firstprivate(@firstprivatization__ZTSA5_f -> %[[PRIVATE]] : !cir.ptr>) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc parallel firstprivate(noCopyArr[1:1]) ; // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST2:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CAST2]] : si32) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[NOCOPYARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "noCopyArr[1:1]"} // CHECK-NEXT: acc.parallel firstprivate(@firstprivatization__ZTSA5_15NoCopyConstruct -> %[[PRIVATE]] : !cir.ptr>) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc serial firstprivate(hasCopyArr[1:1]) ; // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST2:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CAST2]] : si32) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[HASCOPYARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "hasCopyArr[1:1]"} // CHECK-NEXT: acc.serial firstprivate(@firstprivatization__ZTSA5_13CopyConstruct -> %[[PRIVATE]] : !cir.ptr>) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc parallel firstprivate(notDefCtorArr[1:1]) ; // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST2:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CAST2]] : si32) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[NOTDEFCTORARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "notDefCtorArr[1:1]"} // CHECK-NEXT: acc.parallel firstprivate(@firstprivatization__ZTSA5_14NonDefaultCtor -> %[[PRIVATE]] : !cir.ptr>) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc parallel firstprivate(dtorArr[1:1]) ; // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST2:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CAST2]] : si32) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE:.*]] = acc.firstprivate varPtr(%[[DTORARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "dtorArr[1:1]"} // CHECK-NEXT: acc.parallel firstprivate(@firstprivatization__ZTSA5_7HasDtor -> %[[PRIVATE]] : !cir.ptr>) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc #pragma acc parallel firstprivate(someIntArr[1:1], someFloatArr[1:1], noCopyArr[1:1], hasCopyArr[1:1], notDefCtorArr[1:1], dtorArr[1:1]) ; // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST2:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CAST2]] : si32) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE1:.*]] = acc.firstprivate varPtr(%[[INTARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "someIntArr[1:1]"} // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST2:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CAST2]] : si32) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE2:.*]] = acc.firstprivate varPtr(%[[FLOATARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "someFloatArr[1:1]"} // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST2:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CAST2]] : si32) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE3:.*]] = acc.firstprivate varPtr(%[[NOCOPYARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "noCopyArr[1:1]"} // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST2:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CAST2]] : si32) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE4:.*]] = acc.firstprivate varPtr(%[[HASCOPYARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "hasCopyArr[1:1]"} // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST2:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CAST2]] : si32) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE5:.*]] = acc.firstprivate varPtr(%[[NOTDEFCTORARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "notDefCtorArr[1:1]"} // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ONE:.*]] = cir.const #cir.int<1> : !s32i // CHECK-NEXT: %[[ONE_CAST2:.*]] = builtin.unrealized_conversion_cast %[[ONE]] : !s32i to si32 // CHECK-NEXT: %[[ZERO_CONST:.*]] = arith.constant 0 // CHECK-NEXT: %[[ONE_CONST2:.*]] = arith.constant 1 // CHECK-NEXT: %[[BOUNDS:.*]] = acc.bounds lowerbound(%[[ONE_CAST]] : si32) extent(%[[ONE_CAST2]] : si32) stride(%[[ONE_CONST2]] : i64) startIdx(%[[ZERO_CONST]] : i64) // CHECK-NEXT: %[[PRIVATE6:.*]] = acc.firstprivate varPtr(%[[DTORARR]] : !cir.ptr>) bounds(%[[BOUNDS]]) -> !cir.ptr> {name = "dtorArr[1:1]"} // CHECK-NEXT: acc.parallel firstprivate(@firstprivatization__ZTSA5_i -> %[[PRIVATE1]] : !cir.ptr>, // CHECK-SAME: @firstprivatization__ZTSA5_f -> %[[PRIVATE2]] : !cir.ptr>, // CHECK-SAME: @firstprivatization__ZTSA5_15NoCopyConstruct -> %[[PRIVATE3]] : !cir.ptr>, // CHECK-SAME: @firstprivatization__ZTSA5_13CopyConstruct -> %[[PRIVATE4]] : !cir.ptr>, // CHECK-SAME: @firstprivatization__ZTSA5_14NonDefaultCtor -> %[[PRIVATE5]] : !cir.ptr>, // CHECK-SAME: @firstprivatization__ZTSA5_7HasDtor -> %[[PRIVATE6]] : !cir.ptr>) // CHECK-NEXT: acc.yield // CHECK-NEXT: } loc }