I'm not sure, but assume the Inputs directories could be sunk to a common parent (test/Inputs rather than test/CodeGen/Inputs, etc)
On Thu, May 19, 2016 at 5:38 PM, Reid Kleckner via cfe-commits < cfe-commits@lists.llvm.org> wrote: > Author: rnk > Date: Thu May 19 19:38:25 2016 > New Revision: 270164 > > URL: http://llvm.org/viewvc/llvm-project?rev=270164&view=rev > Log: > Avoid depending on test inputes that aren't in Inputs > > Some people have weird CI systems that run each test subdirectory > independently without access to other parallel trees. > > Unfortunately, this means we have to suffer some duplication until Art > can sort out how to share these types. > > Added: > cfe/trunk/test/CodeGenCUDA/Inputs/cuda-initializers.h > cfe/trunk/test/SemaCUDA/Inputs/cuda-initializers.h > Modified: > cfe/trunk/test/CodeGenCUDA/device-var-init.cu > cfe/trunk/test/SemaCUDA/device-var-init.cu > > Added: cfe/trunk/test/CodeGenCUDA/Inputs/cuda-initializers.h > URL: > http://llvm.org/viewvc/llvm-project/cfe/trunk/test/CodeGenCUDA/Inputs/cuda-initializers.h?rev=270164&view=auto > > ============================================================================== > --- cfe/trunk/test/CodeGenCUDA/Inputs/cuda-initializers.h (added) > +++ cfe/trunk/test/CodeGenCUDA/Inputs/cuda-initializers.h Thu May 19 > 19:38:25 2016 > @@ -0,0 +1,145 @@ > +// CUDA struct types with interesting initialization properties. > +// Keep in sync with ../SemaCUDA/Inputs/cuda-initializers.h. > + > +// Base classes with different initializer variants. > + > +// trivial constructor -- allowed > +struct T { > + int t; > +}; > + > +// empty constructor > +struct EC { > + int ec; > + __device__ EC() {} // -- allowed > + __device__ EC(int) {} // -- not allowed > +}; > + > +// empty destructor > +struct ED { > + __device__ ~ED() {} // -- allowed > +}; > + > +struct ECD { > + __device__ ECD() {} // -- allowed > + __device__ ~ECD() {} // -- allowed > +}; > + > +// empty templated constructor -- allowed with no arguments > +struct ETC { > + template <typename... T> __device__ ETC(T...) {} > +}; > + > +// undefined constructor -- not allowed > +struct UC { > + int uc; > + __device__ UC(); > +}; > + > +// undefined destructor -- not allowed > +struct UD { > + int ud; > + __device__ ~UD(); > +}; > + > +// empty constructor w/ initializer list -- not allowed > +struct ECI { > + int eci; > + __device__ ECI() : eci(1) {} > +}; > + > +// non-empty constructor -- not allowed > +struct NEC { > + int nec; > + __device__ NEC() { nec = 1; } > +}; > + > +// non-empty destructor -- not allowed > +struct NED { > + int ned; > + __device__ ~NED() { ned = 1; } > +}; > + > +// no-constructor, virtual method -- not allowed > +struct NCV { > + int ncv; > + __device__ virtual void vm() {} > +}; > + > +// virtual destructor -- not allowed. > +struct VD { > + __device__ virtual ~VD() {} > +}; > + > +// dynamic in-class field initializer -- not allowed > +__device__ int f(); > +struct NCF { > + int ncf = f(); > +}; > + > +// static in-class field initializer. NVCC does not allow it, but > +// clang generates static initializer for this, so we'll accept it. > +// We still can't use it on __shared__ vars as they don't allow *any* > +// initializers. > +struct NCFS { > + int ncfs = 3; > +}; > + > +// undefined templated constructor -- not allowed > +struct UTC { > + template <typename... T> __device__ UTC(T...); > +}; > + > +// non-empty templated constructor -- not allowed > +struct NETC { > + int netc; > + template <typename... T> __device__ NETC(T...) { netc = 1; } > +}; > + > +// Regular base class -- allowed > +struct T_B_T : T {}; > + > +// Incapsulated object of allowed class -- allowed > +struct T_F_T { > + T t; > +}; > + > +// array of allowed objects -- allowed > +struct T_FA_T { > + T t[2]; > +}; > + > + > +// Calling empty base class initializer is OK > +struct EC_I_EC : EC { > + __device__ EC_I_EC() : EC() {} > +}; > + > +// .. though passing arguments is not allowed. > +struct EC_I_EC1 : EC { > + __device__ EC_I_EC1() : EC(1) {} > +}; > + > +// Virtual base class -- not allowed > +struct T_V_T : virtual T {}; > + > +// Inherited from or incapsulated class with non-empty constructor -- > +// not allowed > +struct T_B_NEC : NEC {}; > +struct T_F_NEC { > + NEC nec; > +}; > +struct T_FA_NEC { > + NEC nec[2]; > +}; > + > + > +// Inherited from or incapsulated class with non-empty desstructor -- > +// not allowed > +struct T_B_NED : NED {}; > +struct T_F_NED { > + NED ned; > +}; > +struct T_FA_NED { > + NED ned[2]; > +}; > > Modified: cfe/trunk/test/CodeGenCUDA/device-var-init.cu > URL: > http://llvm.org/viewvc/llvm-project/cfe/trunk/test/CodeGenCUDA/device-var-init.cu?rev=270164&r1=270163&r2=270164&view=diff > > ============================================================================== > --- cfe/trunk/test/CodeGenCUDA/device-var-init.cu (original) > +++ cfe/trunk/test/CodeGenCUDA/device-var-init.cu Thu May 19 19:38:25 2016 > @@ -10,100 +10,8 @@ > #include "Inputs/cuda.h" > #endif > > -// Base classes with different initializer variants. > - > -// trivial constructor -- allowed > -struct T { > - int t; > -}; > - > -// empty constructor > -struct EC { > - int ec; > - __device__ EC() {} // -- allowed > - __device__ EC(int) {} // -- not allowed > -}; > - > -// empty destructor > -struct ED { > - __device__ ~ED() {} // -- allowed > -}; > - > -struct ECD { > - __device__ ECD() {} // -- allowed > - __device__ ~ECD() {} // -- allowed > -}; > - > -// empty templated constructor -- allowed with no arguments > -struct ETC { > - template <typename... T> __device__ ETC(T...) {} > -}; > - > -// undefined constructor -- not allowed > -struct UC { > - int uc; > - __device__ UC(); > -}; > - > -// undefined destructor -- not allowed > -struct UD { > - int ud; > - __device__ ~UD(); > -}; > - > -// empty constructor w/ initializer list -- not allowed > -struct ECI { > - int eci; > - __device__ ECI() : eci(1) {} > -}; > - > -// non-empty constructor -- not allowed > -struct NEC { > - int nec; > - __device__ NEC() { nec = 1; } > -}; > - > -// non-empty destructor -- not allowed > -struct NED { > - int ned; > - __device__ ~NED() { ned = 1; } > -}; > - > -// no-constructor, virtual method -- not allowed > -struct NCV { > - int ncv; > - __device__ virtual void vm() {} > -}; > - > -// virtual destructor -- not allowed. > -struct VD { > - __device__ virtual ~VD() {} > -}; > - > -// dynamic in-class field initializer -- not allowed > -__device__ int f(); > -struct NCF { > - int ncf = f(); > -}; > - > -// static in-class field initializer. NVCC does not allow it, but > -// clang generates static initializer for this, so we'll accept it. > -// We still can't use it on __shared__ vars as they don't allow *any* > -// initializers. > -struct NCFS { > - int ncfs = 3; > -}; > - > -// undefined templated constructor -- not allowed > -struct UTC { > - template <typename... T> __device__ UTC(T...); > -}; > - > -// non-empty templated constructor -- not allowed > -struct NETC { > - int netc; > - template <typename... T> __device__ NETC(T...) { netc = 1; } > -}; > +// Use the types we share with Sema tests. > +#include "Inputs/cuda-initializers.h" > > __device__ int d_v; > // CHECK: @d_v = addrspace(1) externally_initialized global i32 0, > @@ -115,6 +23,7 @@ __constant__ int c_v; > __device__ int d_v_i = 1; > // CHECK: @d_v_i = addrspace(1) externally_initialized global i32 1, > > +// trivial constructor -- allowed > __device__ T d_t; > // CHECK: @d_t = addrspace(1) externally_initialized global %struct.T > zeroinitializer > __shared__ T s_t; > @@ -127,6 +36,7 @@ __device__ T d_t_i = {2}; > __constant__ T c_t_i = {2}; > // CHECK: @c_t_i = addrspace(4) externally_initialized global %struct.T { > i32 2 }, > > +// empty constructor > __device__ EC d_ec; > // CHECK: @d_ec = addrspace(1) externally_initialized global %struct.EC > zeroinitializer, > __shared__ EC s_ec; > @@ -134,6 +44,7 @@ __shared__ EC s_ec; > __constant__ EC c_ec; > // CHECK: @c_ec = addrspace(4) externally_initialized global %struct.EC > zeroinitializer, > > +// empty destructor > __device__ ED d_ed; > // CHECK: @d_ed = addrspace(1) externally_initialized global %struct.ED > zeroinitializer, > __shared__ ED s_ed; > @@ -148,6 +59,7 @@ __shared__ ECD s_ecd; > __constant__ ECD c_ecd; > // CHECK: @c_ecd = addrspace(4) externally_initialized global %struct.ECD > zeroinitializer, > > +// empty templated constructor -- allowed with no arguments > __device__ ETC d_etc; > // CHECK: @d_etc = addrspace(1) externally_initialized global %struct.ETC > zeroinitializer, > __shared__ ETC s_etc; > @@ -161,7 +73,6 @@ __constant__ NCFS c_ncfs; > // CHECK: @c_ncfs = addrspace(4) externally_initialized global > %struct.NCFS { i32 3 } > > // Regular base class -- allowed > -struct T_B_T : T {}; > __device__ T_B_T d_t_b_t; > // CHECK: @d_t_b_t = addrspace(1) externally_initialized global > %struct.T_B_T zeroinitializer, > __shared__ T_B_T s_t_b_t; > @@ -170,9 +81,6 @@ __constant__ T_B_T c_t_b_t; > // CHECK: @c_t_b_t = addrspace(4) externally_initialized global > %struct.T_B_T zeroinitializer, > > // Incapsulated object of allowed class -- allowed > -struct T_F_T { > - T t; > -}; > __device__ T_F_T d_t_f_t; > // CHECK: @d_t_f_t = addrspace(1) externally_initialized global > %struct.T_F_T zeroinitializer, > __shared__ T_F_T s_t_f_t; > @@ -181,9 +89,6 @@ __constant__ T_F_T c_t_f_t; > // CHECK: @c_t_f_t = addrspace(4) externally_initialized global > %struct.T_F_T zeroinitializer, > > // array of allowed objects -- allowed > -struct T_FA_T { > - T t[2]; > -}; > __device__ T_FA_T d_t_fa_t; > // CHECK: @d_t_fa_t = addrspace(1) externally_initialized global > %struct.T_FA_T zeroinitializer, > __shared__ T_FA_T s_t_fa_t; > @@ -193,9 +98,6 @@ __constant__ T_FA_T c_t_fa_t; > > > // Calling empty base class initializer is OK > -struct EC_I_EC : EC { > - __device__ EC_I_EC() : EC() {} > -}; > __device__ EC_I_EC d_ec_i_ec; > // CHECK: @d_ec_i_ec = addrspace(1) externally_initialized global > %struct.EC_I_EC zeroinitializer, > __shared__ EC_I_EC s_ec_i_ec; > @@ -203,35 +105,6 @@ __shared__ EC_I_EC s_ec_i_ec; > __constant__ EC_I_EC c_ec_i_ec; > // CHECK: @c_ec_i_ec = addrspace(4) externally_initialized global > %struct.EC_I_EC zeroinitializer, > > -// .. though passing arguments is not allowed. > -struct EC_I_EC1 : EC { > - __device__ EC_I_EC1() : EC(1) {} > -}; > - > -// Virtual base class -- not allowed > -struct T_V_T : virtual T {}; > - > -// Inherited from or incapsulated class with non-empty constructor -- > -// not allowed > -struct T_B_NEC : NEC {}; > -struct T_F_NEC { > - NEC nec; > -}; > -struct T_FA_NEC { > - NEC nec[2]; > -}; > - > - > -// Inherited from or incapsulated class with non-empty desstructor -- > -// not allowed > -struct T_B_NED : NED {}; > -struct T_F_NED { > - NED ned; > -}; > -struct T_FA_NED { > - NED ned[2]; > -}; > - > // We should not emit global initializers for device-side variables. > // CHECK-NOT: @__cxx_global_var_init > > @@ -249,14 +122,20 @@ __device__ void df() { > ETC etc; > // CHECK: call void @_ZN3ETCC1IJEEEDpT_(%struct.ETC* %etc) > UC uc; > + // undefined constructor -- not allowed > // CHECK: call void @_ZN2UCC1Ev(%struct.UC* %uc) > UD ud; > + // undefined destructor -- not allowed > // CHECK-NOT: call > ECI eci; > + // empty constructor w/ initializer list -- not allowed > // CHECK: call void @_ZN3ECIC1Ev(%struct.ECI* %eci) > NEC nec; > + // non-empty constructor -- not allowed > // CHECK: call void @_ZN3NECC1Ev(%struct.NEC* %nec) > + // non-empty destructor -- not allowed > NED ned; > + // no-constructor, virtual method -- not allowed > // CHECK: call void @_ZN3NCVC1Ev(%struct.NCV* %ncv) > NCV ncv; > // CHECK-NOT: call > > Added: cfe/trunk/test/SemaCUDA/Inputs/cuda-initializers.h > URL: > http://llvm.org/viewvc/llvm-project/cfe/trunk/test/SemaCUDA/Inputs/cuda-initializers.h?rev=270164&view=auto > > ============================================================================== > --- cfe/trunk/test/SemaCUDA/Inputs/cuda-initializers.h (added) > +++ cfe/trunk/test/SemaCUDA/Inputs/cuda-initializers.h Thu May 19 19:38:25 > 2016 > @@ -0,0 +1,145 @@ > +// CUDA struct types with interesting initialization properties. > +// Keep in sync with ../CodeGenCUDA/Inputs/cuda-initializers.h. > + > +// Base classes with different initializer variants. > + > +// trivial constructor -- allowed > +struct T { > + int t; > +}; > + > +// empty constructor > +struct EC { > + int ec; > + __device__ EC() {} // -- allowed > + __device__ EC(int) {} // -- not allowed > +}; > + > +// empty destructor > +struct ED { > + __device__ ~ED() {} // -- allowed > +}; > + > +struct ECD { > + __device__ ECD() {} // -- allowed > + __device__ ~ECD() {} // -- allowed > +}; > + > +// empty templated constructor -- allowed with no arguments > +struct ETC { > + template <typename... T> __device__ ETC(T...) {} > +}; > + > +// undefined constructor -- not allowed > +struct UC { > + int uc; > + __device__ UC(); > +}; > + > +// undefined destructor -- not allowed > +struct UD { > + int ud; > + __device__ ~UD(); > +}; > + > +// empty constructor w/ initializer list -- not allowed > +struct ECI { > + int eci; > + __device__ ECI() : eci(1) {} > +}; > + > +// non-empty constructor -- not allowed > +struct NEC { > + int nec; > + __device__ NEC() { nec = 1; } > +}; > + > +// non-empty destructor -- not allowed > +struct NED { > + int ned; > + __device__ ~NED() { ned = 1; } > +}; > + > +// no-constructor, virtual method -- not allowed > +struct NCV { > + int ncv; > + __device__ virtual void vm() {} > +}; > + > +// virtual destructor -- not allowed. > +struct VD { > + __device__ virtual ~VD() {} > +}; > + > +// dynamic in-class field initializer -- not allowed > +__device__ int f(); > +struct NCF { > + int ncf = f(); > +}; > + > +// static in-class field initializer. NVCC does not allow it, but > +// clang generates static initializer for this, so we'll accept it. > +// We still can't use it on __shared__ vars as they don't allow *any* > +// initializers. > +struct NCFS { > + int ncfs = 3; > +}; > + > +// undefined templated constructor -- not allowed > +struct UTC { > + template <typename... T> __device__ UTC(T...); > +}; > + > +// non-empty templated constructor -- not allowed > +struct NETC { > + int netc; > + template <typename... T> __device__ NETC(T...) { netc = 1; } > +}; > + > +// Regular base class -- allowed > +struct T_B_T : T {}; > + > +// Incapsulated object of allowed class -- allowed > +struct T_F_T { > + T t; > +}; > + > +// array of allowed objects -- allowed > +struct T_FA_T { > + T t[2]; > +}; > + > + > +// Calling empty base class initializer is OK > +struct EC_I_EC : EC { > + __device__ EC_I_EC() : EC() {} > +}; > + > +// .. though passing arguments is not allowed. > +struct EC_I_EC1 : EC { > + __device__ EC_I_EC1() : EC(1) {} > +}; > + > +// Virtual base class -- not allowed > +struct T_V_T : virtual T {}; > + > +// Inherited from or incapsulated class with non-empty constructor -- > +// not allowed > +struct T_B_NEC : NEC {}; > +struct T_F_NEC { > + NEC nec; > +}; > +struct T_FA_NEC { > + NEC nec[2]; > +}; > + > + > +// Inherited from or incapsulated class with non-empty desstructor -- > +// not allowed > +struct T_B_NED : NED {}; > +struct T_F_NED { > + NED ned; > +}; > +struct T_FA_NED { > + NED ned[2]; > +}; > > Modified: cfe/trunk/test/SemaCUDA/device-var-init.cu > URL: > http://llvm.org/viewvc/llvm-project/cfe/trunk/test/SemaCUDA/device-var-init.cu?rev=270164&r1=270163&r2=270164&view=diff > > ============================================================================== > --- cfe/trunk/test/SemaCUDA/device-var-init.cu (original) > +++ cfe/trunk/test/SemaCUDA/device-var-init.cu Thu May 19 19:38:25 2016 > @@ -3,12 +3,14 @@ > // Make sure we don't allow dynamic initialization for device > // variables, but accept empty constructors allowed by CUDA. > > -// RUN: %clang_cc1 -triple nvptx64-nvidia-cuda -fcuda-is-device > -std=c++11 \ > -// RUN: -I %S/.. -fsyntax-only -verify -o /dev/null %s > +// RUN: %clang_cc1 -verify %s -triple nvptx64-nvidia-cuda > -fcuda-is-device -std=c++11 %s > > -// Counterpart in CodeGen covers valid cases that pass Sema > -// checks. Here we'll only cover cases that trigger errors. > -#include "CodeGenCUDA/device-var-init.cu" > +#ifdef __clang__ > +#include "Inputs/cuda.h" > +#endif > + > +// Use the types we share with CodeGen tests. > +#include "Inputs/cuda-initializers.h" > > __shared__ int s_v_i = 1; > // expected-error@-1 {{initialization is not supported for __shared__ > variables.}} > > > _______________________________________________ > cfe-commits mailing list > cfe-commits@lists.llvm.org > http://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits >
_______________________________________________ cfe-commits mailing list cfe-commits@lists.llvm.org http://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
