head	1.1;
branch	1.1.1;
access;
symbols
	netbsd-11-0-RC6:1.1.1.2
	netbsd-11-0-RC5:1.1.1.2
	netbsd-11-0-RC4:1.1.1.2
	netbsd-11-0-RC3:1.1.1.2
	netbsd-11-0-RC2:1.1.1.2
	netbsd-11-0-RC1:1.1.1.2
	perseant-exfatfs-base-20250801:1.1.1.2
	netbsd-11:1.1.1.2.0.10
	netbsd-11-base:1.1.1.2
	netbsd-10-1-RELEASE:1.1.1.2
	perseant-exfatfs-base-20240630:1.1.1.2
	perseant-exfatfs:1.1.1.2.0.8
	perseant-exfatfs-base:1.1.1.2
	netbsd-9-4-RELEASE:1.1.1.1
	netbsd-10-0-RELEASE:1.1.1.2
	netbsd-10-0-RC6:1.1.1.2
	netbsd-10-0-RC5:1.1.1.2
	netbsd-10-0-RC4:1.1.1.2
	netbsd-10-0-RC3:1.1.1.2
	netbsd-10-0-RC2:1.1.1.2
	netbsd-10-0-RC1:1.1.1.2
	netbsd-10:1.1.1.2.0.6
	netbsd-10-base:1.1.1.2
	netbsd-9-3-RELEASE:1.1.1.1
	cjep_sun2x:1.1.1.2.0.4
	cjep_sun2x-base:1.1.1.2
	cjep_staticlib_x-base1:1.1.1.2
	netbsd-9-2-RELEASE:1.1.1.1
	cjep_staticlib_x:1.1.1.2.0.2
	cjep_staticlib_x-base:1.1.1.2
	netbsd-9-1-RELEASE:1.1.1.1
	phil-wifi-20200421:1.1.1.2
	phil-wifi-20200411:1.1.1.2
	phil-wifi-20200406:1.1.1.2
	netbsd-9-0-RELEASE:1.1.1.1
	netbsd-9-0-RC2:1.1.1.1
	netbsd-9-0-RC1:1.1.1.1
	netbsd-9:1.1.1.1.0.6
	netbsd-9-base:1.1.1.1
	phil-wifi:1.1.1.1.0.4
	phil-wifi-20190609:1.1.1.1
	pgoyette-compat-merge-20190127:1.1.1.1.2.2
	pgoyette-compat-20190127:1.1.1.1
	pgoyette-compat-20190118:1.1.1.1
	pgoyette-compat-1226:1.1.1.1
	pgoyette-compat-1126:1.1.1.1
	pgoyette-compat-1020:1.1.1.1
	pgoyette-compat-0930:1.1.1.1
	pgoyette-compat-0906:1.1.1.1
	pgoyette-compat:1.1.1.1.0.2
	pgoyette-compat-0728:1.1.1.1
	clang-337282:1.1.1.1
	LLVM:1.1.1;
locks; strict;
comment	@// @;


1.1
date	2018.07.17.18.31.41;	author joerg;	state Exp;
branches
	1.1.1.1;
next	;
commitid	wDzL46ALjrCZgwKA;

1.1.1.1
date	2018.07.17.18.31.41;	author joerg;	state Exp;
branches
	1.1.1.1.2.1
	1.1.1.1.4.1;
next	1.1.1.2;
commitid	wDzL46ALjrCZgwKA;

1.1.1.2
date	2019.11.13.22.22.43;	author joerg;	state dead;
branches;
next	;
commitid	QD8YATxuNG34YJKB;

1.1.1.1.2.1
date	2018.07.17.18.31.41;	author pgoyette;	state dead;
branches;
next	1.1.1.1.2.2;
commitid	1UP1xAIUxv1ZgRLA;

1.1.1.1.2.2
date	2018.07.28.04.34.05;	author pgoyette;	state Exp;
branches;
next	;
commitid	1UP1xAIUxv1ZgRLA;

1.1.1.1.4.1
date	2018.07.17.18.31.41;	author christos;	state dead;
branches;
next	1.1.1.1.4.2;
commitid	jtc8rnCzWiEEHGqB;

1.1.1.1.4.2
date	2019.06.10.21.46.26;	author christos;	state Exp;
branches;
next	1.1.1.1.4.3;
commitid	jtc8rnCzWiEEHGqB;

1.1.1.1.4.3
date	2020.04.13.07.50.11;	author martin;	state dead;
branches;
next	;
commitid	X01YhRUPVUDaec4C;


desc
@@


1.1
log
@Initial revision
@
text
@// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-ppc-host.bc
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple nvptx64-nvidia-cuda -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck %s
// expected-no-diagnostics

int foo(int &a) { return a; }

int bar() {
  int a;
  return foo(a);
}

// CHECK: define weak void @@__omp_offloading_{{.*}}maini1{{.*}}_l[[@@LINE+5]](i32* dereferenceable{{.*}})
// CHECK-NOT: @@__kmpc_data_sharing_push_stack

int maini1() {
  int a;
#pragma omp target parallel map(from:a)
  {
    int b;
    a = foo(b) + bar();
  }
  return a;
}

// parallel region
// CHECK: define {{.*}}void @@{{.*}}(i32* noalias {{.*}}, i32* noalias {{.*}}, i32* dereferenceable{{.*}})
// CHECK-NOT: call i8* @@__kmpc_data_sharing_push_stack(
// CHECK: [[B_ADDR:%.+]] = alloca i32,
// CHECK: call {{.*}}[[FOO:@@.*foo.*]](i32* dereferenceable{{.*}} [[B_ADDR]])
// CHECK: call {{.*}}[[BAR:@@.*bar.*]]()
// CHECK-NOT: call void @@__kmpc_data_sharing_pop_stack(
// CHECK: ret void

// CHECK: define {{.*}}[[FOO]](i32* dereferenceable{{.*}})
// CHECK-NOT: @@__kmpc_data_sharing_push_stack

// CHECK: define {{.*}}[[BAR]]()
// CHECK: [[RES:%.+]] = call i8* @@__kmpc_data_sharing_push_stack(i64 4, i16 0)
// CHECK: [[GLOBALS:%.+]] = bitcast i8* [[RES]] to [[GLOBAL_ST:%struct[.].*]]*
// CHECK: [[A_ADDR:%.+]] = getelementptr inbounds [[GLOBAL_ST]], [[GLOBAL_ST]]* [[GLOBALS]], i{{[0-9]+}} 0, i{{[0-9]+}} 0
// CHECK: call {{.*}}[[FOO]](i32* dereferenceable{{.*}} [[A_ADDR]])
// CHECK: call void @@__kmpc_data_sharing_pop_stack(i8* [[RES]])
// CHECK: ret i32
@


1.1.1.1
log
@Import clang r337282 from trunk
@
text
@@


1.1.1.2
log
@Mark old LLVM instance as dead.
@
text
@@


1.1.1.1.4.1
log
@file declare_target_codegen_globalization.cpp was added on branch phil-wifi on 2019-06-10 21:46:26 +0000
@
text
@d1 43
@


1.1.1.1.4.2
log
@Sync with HEAD
@
text
@a0 43
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-ppc-host.bc
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple nvptx64-nvidia-cuda -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck %s
// expected-no-diagnostics

int foo(int &a) { return a; }

int bar() {
  int a;
  return foo(a);
}

// CHECK: define weak void @@__omp_offloading_{{.*}}maini1{{.*}}_l[[@@LINE+5]](i32* dereferenceable{{.*}})
// CHECK-NOT: @@__kmpc_data_sharing_push_stack

int maini1() {
  int a;
#pragma omp target parallel map(from:a)
  {
    int b;
    a = foo(b) + bar();
  }
  return a;
}

// parallel region
// CHECK: define {{.*}}void @@{{.*}}(i32* noalias {{.*}}, i32* noalias {{.*}}, i32* dereferenceable{{.*}})
// CHECK-NOT: call i8* @@__kmpc_data_sharing_push_stack(
// CHECK: [[B_ADDR:%.+]] = alloca i32,
// CHECK: call {{.*}}[[FOO:@@.*foo.*]](i32* dereferenceable{{.*}} [[B_ADDR]])
// CHECK: call {{.*}}[[BAR:@@.*bar.*]]()
// CHECK-NOT: call void @@__kmpc_data_sharing_pop_stack(
// CHECK: ret void

// CHECK: define {{.*}}[[FOO]](i32* dereferenceable{{.*}})
// CHECK-NOT: @@__kmpc_data_sharing_push_stack

// CHECK: define {{.*}}[[BAR]]()
// CHECK: [[RES:%.+]] = call i8* @@__kmpc_data_sharing_push_stack(i64 4, i16 0)
// CHECK: [[GLOBALS:%.+]] = bitcast i8* [[RES]] to [[GLOBAL_ST:%struct[.].*]]*
// CHECK: [[A_ADDR:%.+]] = getelementptr inbounds [[GLOBAL_ST]], [[GLOBAL_ST]]* [[GLOBALS]], i{{[0-9]+}} 0, i{{[0-9]+}} 0
// CHECK: call {{.*}}[[FOO]](i32* dereferenceable{{.*}} [[A_ADDR]])
// CHECK: call void @@__kmpc_data_sharing_pop_stack(i8* [[RES]])
// CHECK: ret i32
@


1.1.1.1.4.3
log
@Mostly merge changes from HEAD upto 20200411
@
text
@@


1.1.1.1.2.1
log
@file declare_target_codegen_globalization.cpp was added on branch pgoyette-compat on 2018-07-28 04:34:05 +0000
@
text
@d1 43
@


1.1.1.1.2.2
log
@Sync with HEAD
@
text
@a0 43
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-ppc-host.bc
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple nvptx64-nvidia-cuda -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck %s
// expected-no-diagnostics

int foo(int &a) { return a; }

int bar() {
  int a;
  return foo(a);
}

// CHECK: define weak void @@__omp_offloading_{{.*}}maini1{{.*}}_l[[@@LINE+5]](i32* dereferenceable{{.*}})
// CHECK-NOT: @@__kmpc_data_sharing_push_stack

int maini1() {
  int a;
#pragma omp target parallel map(from:a)
  {
    int b;
    a = foo(b) + bar();
  }
  return a;
}

// parallel region
// CHECK: define {{.*}}void @@{{.*}}(i32* noalias {{.*}}, i32* noalias {{.*}}, i32* dereferenceable{{.*}})
// CHECK-NOT: call i8* @@__kmpc_data_sharing_push_stack(
// CHECK: [[B_ADDR:%.+]] = alloca i32,
// CHECK: call {{.*}}[[FOO:@@.*foo.*]](i32* dereferenceable{{.*}} [[B_ADDR]])
// CHECK: call {{.*}}[[BAR:@@.*bar.*]]()
// CHECK-NOT: call void @@__kmpc_data_sharing_pop_stack(
// CHECK: ret void

// CHECK: define {{.*}}[[FOO]](i32* dereferenceable{{.*}})
// CHECK-NOT: @@__kmpc_data_sharing_push_stack

// CHECK: define {{.*}}[[BAR]]()
// CHECK: [[RES:%.+]] = call i8* @@__kmpc_data_sharing_push_stack(i64 4, i16 0)
// CHECK: [[GLOBALS:%.+]] = bitcast i8* [[RES]] to [[GLOBAL_ST:%struct[.].*]]*
// CHECK: [[A_ADDR:%.+]] = getelementptr inbounds [[GLOBAL_ST]], [[GLOBAL_ST]]* [[GLOBALS]], i{{[0-9]+}} 0, i{{[0-9]+}} 0
// CHECK: call {{.*}}[[FOO]](i32* dereferenceable{{.*}} [[A_ADDR]])
// CHECK: call void @@__kmpc_data_sharing_pop_stack(i8* [[RES]])
// CHECK: ret i32
@