1 // RUN: %clang_cc1 -std=c++11 -fsyntax-only -verify %s -Wno-defaulted-function-deleted 2 3 #include "Inputs/cuda.h" 4 5 //------------------------------------------------------------------------------ 6 // Test 1: infer inherited default ctor to be host. 7 8 struct A1_with_host_ctor { A1_with_host_ctorA1_with_host_ctor9 A1_with_host_ctor() {} 10 }; 11 // expected-note@-3 {{candidate constructor (the implicit copy constructor) not viable}} 12 // expected-note@-4 {{candidate constructor (the implicit move constructor) not viable}} 13 14 // The inherited default constructor is inferred to be host, so we'll encounter 15 // an error when calling it from a __device__ function, but not from a __host__ 16 // function. 17 struct B1_with_implicit_default_ctor : A1_with_host_ctor { 18 using A1_with_host_ctor::A1_with_host_ctor; 19 }; 20 21 // expected-note@-4 {{call to __host__ function from __device__}} 22 // expected-note@-5 {{candidate constructor (the implicit copy constructor) not viable}} 23 // expected-note@-6 {{candidate constructor (the implicit move constructor) not viable}} 24 // expected-note@-6 2{{constructor from base class 'A1_with_host_ctor' inherited here}} 25 hostfoo()26void hostfoo() { 27 B1_with_implicit_default_ctor b; 28 } 29 devicefoo()30__device__ void devicefoo() { 31 B1_with_implicit_default_ctor b; // expected-error {{no matching constructor}} 32 } 33 34 //------------------------------------------------------------------------------ 35 // Test 2: infer inherited default ctor to be device. 36 37 struct A2_with_device_ctor { A2_with_device_ctorA2_with_device_ctor38 __device__ A2_with_device_ctor() {} 39 }; 40 // expected-note@-3 {{candidate constructor (the implicit copy constructor) not viable}} 41 // expected-note@-4 {{candidate constructor (the implicit move constructor) not viable}} 42 43 struct B2_with_implicit_default_ctor : A2_with_device_ctor { 44 using A2_with_device_ctor::A2_with_device_ctor; 45 }; 46 47 // expected-note@-4 {{call to __device__ function from __host__}} 48 // expected-note@-5 {{candidate constructor (the implicit copy constructor) not viable}} 49 // expected-note@-6 {{candidate constructor (the implicit move constructor) not viable}} 50 // expected-note@-6 2{{constructor from base class 'A2_with_device_ctor' inherited here}} 51 hostfoo2()52void hostfoo2() { 53 B2_with_implicit_default_ctor b; // expected-error {{no matching constructor}} 54 } 55 devicefoo2()56__device__ void devicefoo2() { 57 B2_with_implicit_default_ctor b; 58 } 59 60 //------------------------------------------------------------------------------ 61 // Test 3: infer inherited copy ctor 62 63 struct A3_with_device_ctors { A3_with_device_ctorsA3_with_device_ctors64 __host__ A3_with_device_ctors() {} A3_with_device_ctorsA3_with_device_ctors65 __device__ A3_with_device_ctors(const A3_with_device_ctors&) {} 66 }; 67 68 struct B3_with_implicit_ctors : A3_with_device_ctors { 69 using A3_with_device_ctors::A3_with_device_ctors; 70 }; 71 // expected-note@-3 2{{call to __device__ function from __host__ function}} 72 // expected-note@-4 {{default constructor}} 73 74 hostfoo3()75void hostfoo3() { 76 B3_with_implicit_ctors b; // this is OK because the inferred inherited default ctor 77 // here is __host__ 78 B3_with_implicit_ctors b2 = b; // expected-error {{no matching constructor}} 79 80 } 81 82 //------------------------------------------------------------------------------ 83 // Test 4: infer inherited default ctor from a field, not a base 84 85 struct A4_with_host_ctor { A4_with_host_ctorA4_with_host_ctor86 A4_with_host_ctor() {} 87 }; 88 89 struct B4_with_inherited_host_ctor : A4_with_host_ctor{ 90 using A4_with_host_ctor::A4_with_host_ctor; 91 }; 92 93 struct C4_with_implicit_default_ctor { 94 B4_with_inherited_host_ctor field; 95 }; 96 97 // expected-note@-4 {{call to __host__ function from __device__}} 98 // expected-note@-5 {{candidate constructor (the implicit copy constructor) not viable}} 99 // expected-note@-6 {{candidate constructor (the implicit move constructor) not viable}} 100 hostfoo4()101void hostfoo4() { 102 C4_with_implicit_default_ctor b; 103 } 104 devicefoo4()105__device__ void devicefoo4() { 106 C4_with_implicit_default_ctor b; // expected-error {{no matching constructor}} 107 } 108 109 //------------------------------------------------------------------------------ 110 // Test 5: inherited copy ctor with non-const param 111 112 struct A5_copy_ctor_constness { A5_copy_ctor_constnessA5_copy_ctor_constness113 __host__ A5_copy_ctor_constness() {} A5_copy_ctor_constnessA5_copy_ctor_constness114 __host__ A5_copy_ctor_constness(A5_copy_ctor_constness&) {} 115 }; 116 117 struct B5_copy_ctor_constness : A5_copy_ctor_constness { 118 using A5_copy_ctor_constness::A5_copy_ctor_constness; 119 }; 120 121 // expected-note@-4 {{candidate constructor (the implicit copy constructor) not viable: call to __host__ function from __device__ function}} 122 // expected-note@-5 {{candidate constructor (the implicit default constructor) not viable}} 123 hostfoo5(B5_copy_ctor_constness & b_arg)124void hostfoo5(B5_copy_ctor_constness& b_arg) { 125 B5_copy_ctor_constness b = b_arg; 126 } 127 devicefoo5(B5_copy_ctor_constness & b_arg)128__device__ void devicefoo5(B5_copy_ctor_constness& b_arg) { 129 B5_copy_ctor_constness b = b_arg; // expected-error {{no matching constructor}} 130 } 131 132 //------------------------------------------------------------------------------ 133 // Test 6: explicitly defaulted ctor 134 135 struct A6_with_device_ctor { A6_with_device_ctorA6_with_device_ctor136 __device__ A6_with_device_ctor() {} 137 }; 138 139 struct B6_with_defaulted_ctor : A6_with_device_ctor { 140 using A6_with_device_ctor::A6_with_device_ctor; 141 __host__ B6_with_defaulted_ctor() = default; 142 }; 143 144 // expected-note@-3 {{explicitly defaulted function was implicitly deleted here}} 145 // expected-note@-6 {{default constructor of 'B6_with_defaulted_ctor' is implicitly deleted because base class 'A6_with_device_ctor' has no default constructor}} 146 hostfoo6()147void hostfoo6() { 148 B6_with_defaulted_ctor b; // expected-error {{call to implicitly-deleted default constructor}} 149 } 150 devicefoo6()151__device__ void devicefoo6() { 152 B6_with_defaulted_ctor b; 153 } 154 155 //------------------------------------------------------------------------------ 156 // Test 7: inherited copy assignment operator 157 158 struct A7_with_copy_assign { A7_with_copy_assignA7_with_copy_assign159 A7_with_copy_assign() {} operator =A7_with_copy_assign160 __device__ A7_with_copy_assign& operator=(const A7_with_copy_assign&) {} 161 }; 162 163 struct B7_with_copy_assign : A7_with_copy_assign { 164 using A7_with_copy_assign::A7_with_copy_assign; 165 }; 166 167 // expected-note@-4 {{candidate function (the implicit copy assignment operator) not viable: call to __device__ function from __host__ function}} 168 // expected-note@-5 {{candidate function (the implicit move assignment operator) not viable: call to __device__ function from __host__ function}} 169 hostfoo7()170void hostfoo7() { 171 B7_with_copy_assign b1, b2; 172 b1 = b2; // expected-error {{no viable overloaded '='}} 173 } 174 175 //------------------------------------------------------------------------------ 176 // Test 8: inherited move assignment operator 177 178 // definitions for std::move 179 namespace std { 180 inline namespace foo { 181 template <class T> struct remove_reference { typedef T type; }; 182 template <class T> struct remove_reference<T&> { typedef T type; }; 183 template <class T> struct remove_reference<T&&> { typedef T type; }; 184 185 template <class T> typename remove_reference<T>::type&& move(T&& t); 186 } 187 } 188 189 struct A8_with_move_assign { A8_with_move_assignA8_with_move_assign190 A8_with_move_assign() {} operator =A8_with_move_assign191 __device__ A8_with_move_assign& operator=(A8_with_move_assign&&) {} operator =A8_with_move_assign192 __device__ A8_with_move_assign& operator=(const A8_with_move_assign&) {} 193 }; 194 195 struct B8_with_move_assign : A8_with_move_assign { 196 using A8_with_move_assign::A8_with_move_assign; 197 }; 198 199 // expected-note@-4 {{candidate function (the implicit copy assignment operator) not viable: call to __device__ function from __host__ function}} 200 // expected-note@-5 {{candidate function (the implicit move assignment operator) not viable: call to __device__ function from __host__ function}} 201 hostfoo8()202void hostfoo8() { 203 B8_with_move_assign b1, b2; 204 b1 = std::move(b2); // expected-error {{no viable overloaded '='}} 205 } 206