text: // RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm-bc %s -o %t-ppc-host.bc
  // RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck %s
  // RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -emit-pch -o %t
  // RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -include-pch %t -o - | FileCheck %s
  
  // RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm-bc %s -o %t-ppc-host.bc
  // RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o -| FileCheck %s --check-prefix SIMD-ONLY
  // RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -emit-pch -o %t
  // RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -include-pch %t -verify -o - | FileCheck %s --check-prefix SIMD-ONLY
  
  // expected-no-diagnostics
  
  // SIMD-ONLY-NOT: {{__kmpc|__tgt}}
  
  // CHECK-NOT: define {{.*}}{{baz1|baz4|maini1|Base|virtual_}}
  // CHECK-DAG: Bake
  // CHECK-NOT: @{{hhh|ggg|fff|eee}} =
  // CHECK-DAG: @aaa = external global i32,
  // CHECK-DAG: @bbb ={{ dso_local | }}global i32 0,
  // CHECK-DAG: weak constant %struct.__tgt_offload_entry { i8* bitcast (i32* @bbb to i8*),
  // CHECK-DAG: @ccc = external global i32,
  // CHECK-DAG: @ddd ={{ dso_local | }}global i32 0,
  // CHECK-DAG: @hhh_decl_tgt_link_ptr = common global i32* null
  // CHECK-DAG: @ggg_decl_tgt_link_ptr = common global i32* null
  // CHECK-DAG: @fff_decl_tgt_link_ptr = common global i32* null
  // CHECK-DAG: @eee_decl_tgt_link_ptr = common global i32* null
  // CHECK-DAG: @{{.*}}maini1{{.*}}aaa = internal global i64 23,
  // CHECK-DAG: @b ={{ dso_local | }}global i32 15,
  // CHECK-DAG: @d ={{ dso_local | }}global i32 0,
  // CHECK-DAG: @c = external global i32,
  // CHECK-DAG: @globals ={{ dso_local | }}global %struct.S zeroinitializer,
  // CHECK-DAG: [[STAT:@.+stat]] = internal global %struct.S zeroinitializer,
  // CHECK-DAG: [[STAT_REF:@.+]] = internal constant %struct.S* [[STAT]]
  // CHECK-DAG: @out_decl_target ={{ dso_local | }}global i32 0,
  // CHECK-DAG: @llvm.used = appending global [6 x i8*] [i8* bitcast (void ()* @__omp_offloading__{{.+}}_globals_l[[@LINE+80]]_ctor to i8*), i8* bitcast (void ()* @__omp_offloading__{{.+}}_stat_l[[@LINE+81]]_ctor to i8*),
  // CHECK-DAG: @llvm.compiler.used = appending global [1 x i8*] [i8* bitcast (%struct.S** [[STAT_REF]] to i8*)],
  
  // CHECK-DAG: define {{.*}}i32 @{{.*}}{{foo|bar|baz2|baz3|FA|f_method}}{{.*}}()
  // CHECK-DAG: define {{.*}}void @{{.*}}TemplateClass{{.*}}(%class.TemplateClass* %{{.*}})
  // CHECK-DAG: define {{.*}}i32 @{{.*}}TemplateClass{{.*}}f_method{{.*}}(%class.TemplateClass* %{{.*}})
  // CHECK-DAG: define {{.*}}void @__omp_offloading__{{.*}}_globals_l[[@LINE+74]]_ctor()
  
  #ifndef HEADER
  #define HEADER
  
  #pragma omp declare target
  extern int bbb;
  #pragma omp end declare target
  #pragma omp declare target
  extern int bbb;
  #pragma omp end declare target
  
  #pragma omp declare target
  extern int aaa;
  int bbb = 0;
  extern int ccc;
  int ddd = 0;
  #pragma omp end declare target
  
  #pragma omp declare target
  extern int bbb;
  #pragma omp end declare target
  
  extern int eee;
  int fff = 0;
  extern int ggg;
  int hhh = 0;
  #pragma omp declare target link(eee, fff, ggg, hhh)
  
  int out_decl_target = 0;
  #pragma omp declare target
  void lambda () {
  #ifdef __cpp_lambdas
    (void)[&] { (void)out_decl_target; };
  #else
  #pragma clang __debug captured
    {
      (void)out_decl_target;
    }
  #endif
  };
  #pragma omp end declare target
  
  template <typename T>
  class TemplateClass {
    T a;
  public:
    TemplateClass() {}
    T f_method() const { return a; }
  };
  
  int foo();
  
  static int baz1() { return 0; }
  
  int baz2();
  
  int baz4() { return 5; }
  
  template <typename T>
  T FA() {
    TemplateClass<T> s;
    return s.f_method();
  }
  
  #pragma omp declare target
  struct S {
    int a;
    S(int a) : a(a) {}
  };
  
  int foo() { return 0; }
  int b = 15;
  int d;
  S globals(d);
  static S stat(d);
  #pragma omp end declare target
  int c;
  
  int bar() { return 1 + foo() + bar() + baz1() + baz2(); }
  
  int maini1() {
    int a;
    static long aa = 32 + bbb + ccc + fff + ggg;
  // CHECK-DAG: define weak void @__omp_offloading_{{.*}}maini1{{.*}}_l[[@LINE+1]](i32* dereferenceable{{.*}}, i64 {{.*}}, i64 {{.*}})
  #pragma omp target map(tofrom \
                         : a, b)
    {
      S s(a);
      static long aaa = 23;
      a = foo() + bar() + b + c + d + aa + aaa + FA<int>();
    }
    return baz4();
  }
  
  int baz3() { return 2 + baz2(); }
  int baz2() {
  // CHECK-DAG: define weak void @__omp_offloading_{{.*}}baz2{{.*}}_l[[@LINE+1]](i64 {{.*}})
  #pragma omp target parallel
    ++c;
    return 2 + baz3();
  }
  
  extern int create() throw();
  
  static __typeof(create) __t_create __attribute__((__weakref__("__create")));
  
  int baz5() {
    bool a;
  // CHECK-DAG: define weak void @__omp_offloading_{{.*}}baz5{{.*}}_l[[@LINE+1]](i64 {{.*}})
  #pragma omp target
    a = __extension__(void *) & __t_create != 0;
    return a;
  }
  
  template <typename T>
  struct Base {
    virtual ~Base() {}
  };
  
  template class Base<int>;
  
  template <typename T>
  struct Bake {
    virtual ~Bake() {}
  };
  
  #pragma omp declare target
  template class Bake<int>;
  #pragma omp end declare target
  
  struct BaseNonT {
    virtual ~BaseNonT() {}
  };
  
  #pragma omp declare target
  struct BakeNonT {
    virtual ~BakeNonT() {}
  };
  #pragma omp end declare target
  
  template <typename T>
  struct B {
    virtual void virtual_foo();
  };
  
  void new_bar() { new B<int>(); }
  
  template <typename T>
  void B<T>::virtual_foo() {
  #pragma omp target
    {}
  }
  
  struct A {
    virtual void emitted() {}
  };
  
  template <typename T>
  struct C : public A {
    virtual void emitted();
  };
  
  template <typename T>
  void C<T>::emitted() {
  #pragma omp target
    {}
  }
  
  int main() {
    A *X = new C<int>();
    X->emitted();
    return 0;
  }
  
  // CHECK-DAG: define {{.*}}void @__omp_offloading_{{.*}}virtual_foo{{.*}}_l[[@LINE-25]]()
  // CHECK-DAG: define {{.*}}void @__omp_offloading_{{.*}}emitted{{.*}}_l[[@LINE-11]]()
  
  // CHECK-DAG: declare extern_weak signext i32 @__create()
  
  // CHECK-NOT: define {{.*}}{{baz1|baz4|maini1|Base|virtual_}}
  
  // CHECK-DAG: !{i32 1, !"aaa", i32 0, i32 {{[0-9]+}}}
  // CHECK-DAG: !{i32 1, !"ccc", i32 0, i32 {{[0-9]+}}}
  // CHECK-DAG: !{{{.+}}virtual_foo
  
  #endif // HEADER
  