This revision was automatically updated to reflect the committed changes.
Closed by commit rL283120: [CUDA] Disallow overloading destructors. (authored
by jlebar).
Changed prior to commit:
https://reviews.llvm.org/D24571?vs=71379&id=73413#toc
Repository:
rL LLVM
https://reviews.llvm.org/D24571
Files:
cfe/trunk/lib/Sema/SemaOverload.cpp
cfe/trunk/test/CodeGenCUDA/function-overload.cu
cfe/trunk/test/SemaCUDA/call-overloaded-destructor.cu
cfe/trunk/test/SemaCUDA/function-overload.cu
cfe/trunk/test/SemaCUDA/no-destructor-overload.cu
Index: cfe/trunk/test/CodeGenCUDA/function-overload.cu
===================================================================
--- cfe/trunk/test/CodeGenCUDA/function-overload.cu
+++ cfe/trunk/test/CodeGenCUDA/function-overload.cu
@@ -16,8 +16,6 @@
struct s_cd_dh {
__host__ s_cd_dh() { x = 11; }
__device__ s_cd_dh() { x = 12; }
- __host__ ~s_cd_dh() { x = 21; }
- __device__ ~s_cd_dh() { x = 22; }
};
struct s_cd_hd {
@@ -38,7 +36,6 @@
// CHECK-BOTH: call void @_ZN7s_cd_hdC1Ev
// CHECK-BOTH: call void @_ZN7s_cd_hdD1Ev(
- // CHECK-BOTH: call void @_ZN7s_cd_dhD1Ev(
}
// CHECK-BOTH: ret void
@@ -56,8 +53,3 @@
// CHECK-BOTH: define linkonce_odr void @_ZN7s_cd_hdD2Ev(
// CHECK-BOTH: store i32 32,
// CHECK-BOTH: ret void
-
-// CHECK-BOTH: define linkonce_odr void @_ZN7s_cd_dhD2Ev(
-// CHECK-HOST: store i32 21,
-// CHECK-DEVICE: store i32 22,
-// CHECK-BOTH: ret void
Index: cfe/trunk/test/SemaCUDA/no-destructor-overload.cu
===================================================================
--- cfe/trunk/test/SemaCUDA/no-destructor-overload.cu
+++ cfe/trunk/test/SemaCUDA/no-destructor-overload.cu
@@ -0,0 +1,33 @@
+// RUN: %clang_cc1 -fsyntax-only -verify %s
+// RUN: %clang_cc1 -fcuda-is-device -fsyntax-only -verify %s
+
+#include "Inputs/cuda.h"
+
+// We don't allow destructors to be overloaded. Making this work would be a
+// giant change to clang, and the use cases seem quite limited.
+
+struct A {
+ ~A() {} // expected-note {{previous declaration is here}}
+ __device__ ~A() {} // expected-error {{destructor cannot be redeclared}}
+};
+
+struct B {
+ __host__ ~B() {} // expected-note {{previous declaration is here}}
+ __host__ __device__ ~B() {} // expected-error {{destructor cannot be redeclared}}
+};
+
+struct C {
+ __host__ __device__ ~C() {} // expected-note {{previous declaration is here}}
+ __host__ ~C() {} // expected-error {{destructor cannot be redeclared}}
+};
+
+struct D {
+ __device__ ~D() {} // expected-note {{previous declaration is here}}
+ __host__ __device__ ~D() {} // expected-error {{destructor cannot be redeclared}}
+};
+
+struct E {
+ __host__ __device__ ~E() {} // expected-note {{previous declaration is here}}
+ __device__ ~E() {} // expected-error {{destructor cannot be redeclared}}
+};
+
Index: cfe/trunk/test/SemaCUDA/function-overload.cu
===================================================================
--- cfe/trunk/test/SemaCUDA/function-overload.cu
+++ cfe/trunk/test/SemaCUDA/function-overload.cu
@@ -210,44 +210,11 @@
__host__ ~d_h() {} // expected-error {{destructor cannot be redeclared}}
};
-// H/D overloading is OK
-struct d_dh {
- __device__ ~d_dh() {}
- __host__ ~d_dh() {}
-};
-
// HD is OK
struct d_hd {
__host__ __device__ ~d_hd() {}
};
-// Mixing H/D and HD is not allowed.
-struct d_dhhd {
- __device__ ~d_dhhd() {}
- __host__ ~d_dhhd() {} // expected-note {{previous declaration is here}}
- __host__ __device__ ~d_dhhd() {} // expected-error {{destructor cannot be redeclared}}
-};
-
-struct d_hhd {
- __host__ ~d_hhd() {} // expected-note {{previous declaration is here}}
- __host__ __device__ ~d_hhd() {} // expected-error {{destructor cannot be redeclared}}
-};
-
-struct d_hdh {
- __host__ __device__ ~d_hdh() {} // expected-note {{previous declaration is here}}
- __host__ ~d_hdh() {} // expected-error {{destructor cannot be redeclared}}
-};
-
-struct d_dhd {
- __device__ ~d_dhd() {} // expected-note {{previous declaration is here}}
- __host__ __device__ ~d_dhd() {} // expected-error {{destructor cannot be redeclared}}
-};
-
-struct d_hdd {
- __host__ __device__ ~d_hdd() {} // expected-note {{previous declaration is here}}
- __device__ ~d_hdd() {} // expected-error {{destructor cannot be redeclared}}
-};
-
// Test overloading of member functions
struct m_h {
void operator delete(void *ptr); // expected-note {{previous declaration is here}}
Index: cfe/trunk/test/SemaCUDA/call-overloaded-destructor.cu
===================================================================
--- cfe/trunk/test/SemaCUDA/call-overloaded-destructor.cu
+++ cfe/trunk/test/SemaCUDA/call-overloaded-destructor.cu
@@ -1,17 +0,0 @@
-// expected-no-diagnostics
-
-// RUN: %clang_cc1 -triple x86_64-unknown-linux-gnu -fsyntax-only -verify %s
-// RUN: %clang_cc1 -triple nvptx64-nvidia-cuda -fsyntax-only -fcuda-is-device -verify %s
-
-#include "Inputs/cuda.h"
-
-struct S {
- __host__ ~S() {}
- __device__ ~S() {}
-};
-
-__host__ __device__ void test() {
- S s;
- // This should not crash clang.
- s.~S();
-}
Index: cfe/trunk/lib/Sema/SemaOverload.cpp
===================================================================
--- cfe/trunk/lib/Sema/SemaOverload.cpp
+++ cfe/trunk/lib/Sema/SemaOverload.cpp
@@ -1129,6 +1129,11 @@
}
if (getLangOpts().CUDA && ConsiderCudaAttrs) {
+ // Don't allow overloading of destructors. (In theory we could, but it
+ // would be a giant change to clang.)
+ if (isa<CXXDestructorDecl>(New))
+ return false;
+
CUDAFunctionTarget NewTarget = IdentifyCUDATarget(New),
OldTarget = IdentifyCUDATarget(Old);
if (NewTarget == CFT_InvalidTarget || NewTarget == CFT_Global)
_______________________________________________
cfe-commits mailing list
[email protected]
http://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
