[PATCH] D21912: [CUDA] Don't assume that destructors can't be overloaded.

[Justin Lebar via cfe-commits]

jlebar created this revision.
jlebar added a reviewer: rsmith.
jlebar added subscribers: tra, cfe-commits.

You can overload a destructor in CUDA, and SemaOverload needs to be
tweaked not to crash when it sees an explicit call to an overloaded
destructor.

http://reviews.llvm.org/D21912

Files:
  lib/Sema/SemaOverload.cpp
  test/SemaCUDA/call-overloaded-destructor.cu

Index: test/SemaCUDA/call-overloaded-destructor.cu
===================================================================
--- /dev/null
+++ test/SemaCUDA/call-overloaded-destructor.cu
@@ -0,0 +1,17 @@
+// 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: lib/Sema/SemaOverload.cpp
===================================================================
--- lib/Sema/SemaOverload.cpp
+++ lib/Sema/SemaOverload.cpp
@@ -12407,8 +12407,7 @@
   if (CXXDestructorDecl *DD =
           dyn_cast<CXXDestructorDecl>(TheCall->getMethodDecl())) {
     // a->A::f() doesn't go through the vtable, except in AppleKext mode.
-    bool CallCanBeVirtual = !cast<MemberExpr>(NakedMemExpr)->hasQualifier() ||
-                            getLangOpts().AppleKext;
+    bool CallCanBeVirtual = !MemExpr->hasQualifier() || 
getLangOpts().AppleKext;
     CheckVirtualDtorCall(DD, MemExpr->getLocStart(), /*IsDelete=*/false,
                          CallCanBeVirtual, /*WarnOnNonAbstractTypes=*/true,
                          MemExpr->getMemberLoc());



Index: test/SemaCUDA/call-overloaded-destructor.cu
===================================================================
--- /dev/null
+++ test/SemaCUDA/call-overloaded-destructor.cu
@@ -0,0 +1,17 @@
+// 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: lib/Sema/SemaOverload.cpp
===================================================================
--- lib/Sema/SemaOverload.cpp
+++ lib/Sema/SemaOverload.cpp
@@ -12407,8 +12407,7 @@
   if (CXXDestructorDecl *DD =
           dyn_cast<CXXDestructorDecl>(TheCall->getMethodDecl())) {
     // a->A::f() doesn't go through the vtable, except in AppleKext mode.
-    bool CallCanBeVirtual = !cast<MemberExpr>(NakedMemExpr)->hasQualifier() ||
-                            getLangOpts().AppleKext;
+    bool CallCanBeVirtual = !MemExpr->hasQualifier() || getLangOpts().AppleKext;
     CheckVirtualDtorCall(DD, MemExpr->getLocStart(), /*IsDelete=*/false,
                          CallCanBeVirtual, /*WarnOnNonAbstractTypes=*/true,
                          MemExpr->getMemberLoc());

_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
http://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits