xref: /minix3/external/bsd/llvm/dist/clang/test/SemaCUDA/implicit-copy.cu (revision 0a6a1f1d05b60e214de2f05a7310ddd1f0e590e7)
1*0a6a1f1dSLionel Sambuc // RUN: %clang_cc1 -std=gnu++11 -triple nvptx64-unknown-unknown -fsyntax-only -verify %s
2*0a6a1f1dSLionel Sambuc // RUN: %clang_cc1 -std=gnu++11 -triple nvptx64-unknown-unknown -fcuda-is-device -fsyntax-only -verify %s
3*0a6a1f1dSLionel Sambuc 
4*0a6a1f1dSLionel Sambuc struct CopyableH {
operator =CopyableH5*0a6a1f1dSLionel Sambuc   const CopyableH& operator=(const CopyableH& x) { return *this; }
6*0a6a1f1dSLionel Sambuc };
7*0a6a1f1dSLionel Sambuc struct CopyableD {
operator =CopyableD8*0a6a1f1dSLionel Sambuc   __attribute__((device)) const CopyableD& operator=(const CopyableD x) { return *this; }
9*0a6a1f1dSLionel Sambuc };
10*0a6a1f1dSLionel Sambuc 
11*0a6a1f1dSLionel Sambuc struct SimpleH {
12*0a6a1f1dSLionel Sambuc   CopyableH b;
13*0a6a1f1dSLionel Sambuc };
14*0a6a1f1dSLionel Sambuc // expected-note@-3 2 {{candidate function (the implicit copy assignment operator) not viable: call to __host__ function from __device__ function}}
15*0a6a1f1dSLionel Sambuc // expected-note@-4 2 {{candidate function (the implicit move assignment operator) not viable: call to __host__ function from __device__ function}}
16*0a6a1f1dSLionel Sambuc 
17*0a6a1f1dSLionel Sambuc struct SimpleD {
18*0a6a1f1dSLionel Sambuc   CopyableD b;
19*0a6a1f1dSLionel Sambuc };
20*0a6a1f1dSLionel Sambuc // expected-note@-3 2 {{candidate function (the implicit copy assignment operator) not viable: call to __device__ function from __host__ function}}
21*0a6a1f1dSLionel Sambuc // expected-note@-4 2 {{candidate function (the implicit move assignment operator) not viable: call to __device__ function from __host__ function}}
22*0a6a1f1dSLionel Sambuc 
foo1hh()23*0a6a1f1dSLionel Sambuc void foo1hh() {
24*0a6a1f1dSLionel Sambuc   SimpleH a, b;
25*0a6a1f1dSLionel Sambuc   a = b;
26*0a6a1f1dSLionel Sambuc }
foo1hd()27*0a6a1f1dSLionel Sambuc __attribute__((device)) void foo1hd() {
28*0a6a1f1dSLionel Sambuc   SimpleH a, b;
29*0a6a1f1dSLionel Sambuc   a = b; // expected-error {{no viable overloaded}}
30*0a6a1f1dSLionel Sambuc }
foo1dh()31*0a6a1f1dSLionel Sambuc void foo1dh() {
32*0a6a1f1dSLionel Sambuc   SimpleD a, b;
33*0a6a1f1dSLionel Sambuc   a = b; // expected-error {{no viable overloaded}}
34*0a6a1f1dSLionel Sambuc }
foo1dd()35*0a6a1f1dSLionel Sambuc __attribute__((device)) void foo1dd() {
36*0a6a1f1dSLionel Sambuc   SimpleD a, b;
37*0a6a1f1dSLionel Sambuc   a = b;
38*0a6a1f1dSLionel Sambuc }
39*0a6a1f1dSLionel Sambuc 
foo2hh(SimpleH & a,SimpleH & b)40*0a6a1f1dSLionel Sambuc void foo2hh(SimpleH &a, SimpleH &b) {
41*0a6a1f1dSLionel Sambuc   a = b;
42*0a6a1f1dSLionel Sambuc }
foo2hd(SimpleH & a,SimpleH & b)43*0a6a1f1dSLionel Sambuc __attribute__((device)) void foo2hd(SimpleH &a, SimpleH &b) {
44*0a6a1f1dSLionel Sambuc   a = b; // expected-error {{no viable overloaded}}
45*0a6a1f1dSLionel Sambuc }
foo2dh(SimpleD & a,SimpleD & b)46*0a6a1f1dSLionel Sambuc void foo2dh(SimpleD &a, SimpleD &b) {
47*0a6a1f1dSLionel Sambuc   a = b; // expected-error {{no viable overloaded}}
48*0a6a1f1dSLionel Sambuc }
foo2dd(SimpleD & a,SimpleD & b)49*0a6a1f1dSLionel Sambuc __attribute__((device)) void foo2dd(SimpleD &a, SimpleD &b) {
50*0a6a1f1dSLionel Sambuc   a = b;
51*0a6a1f1dSLionel Sambuc }
52