[PATCH] D47694: [CUDA][HIP] Do not emit type info when compiling for device

Yaxun Liu via Phabricator via cfe-commits cfe-commits at lists.llvm.org
Tue Jun 5 08:15:15 PDT 2018


This revision was automatically updated to reflect the committed changes.
Closed by commit rC334021: [CUDA][HIP] Do not emit type info when compiling for device (authored by yaxunl, committed by ).

Repository:
  rC Clang

https://reviews.llvm.org/D47694

Files:
  lib/CodeGen/CodeGenModule.cpp
  test/CodeGenCUDA/device-vtable.cu


Index: test/CodeGenCUDA/device-vtable.cu
===================================================================
--- test/CodeGenCUDA/device-vtable.cu
+++ test/CodeGenCUDA/device-vtable.cu
@@ -19,15 +19,19 @@
 //CHECK-HOST: @_ZTV1H =
 //CHECK-HOST-SAME: @_ZN1H6methodEv
 //CHECK-DEVICE-NOT: @_ZTV1H =
-
+//CHECK-DEVICE-NOT: @_ZTVN10__cxxabiv117__class_type_infoE
+//CHECK-DEVICE-NOT: @_ZTS1H
+//CHECK-DEVICE-NOT: @_ZTI1H
 struct D  {
    __device__ virtual void method();
 };
 
 //CHECK-DEVICE: @_ZTV1D
 //CHECK-DEVICE-SAME: @_ZN1D6methodEv
 //CHECK-HOST-NOT: @_ZTV1D
-
+//CHECK-DEVICE-NOT: @_ZTVN10__cxxabiv117__class_type_infoE
+//CHECK-DEVICE-NOT: @_ZTS1D
+//CHECK-DEVICE-NOT: @_ZTI1D
 // This is the case with mixed host and device virtual methods.  It's
 // impossible to emit a valid vtable in that case because only host or
 // only device methods would be available during host or device
@@ -45,6 +49,9 @@
 // CHECK-HOST-NOT: @_ZN2HD8d_methodEv
 // CHECK-HOST-SAME: null
 // CHECK-BOTH-SAME: ]
+// CHECK-DEVICE-NOT: @_ZTVN10__cxxabiv117__class_type_infoE
+// CHECK-DEVICE-NOT: @_ZTS2HD
+// CHECK-DEVICE-NOT: @_ZTI2HD
 
 void H::method() {}
 //CHECK-HOST: define void @_ZN1H6methodEv
Index: lib/CodeGen/CodeGenModule.cpp
===================================================================
--- lib/CodeGen/CodeGenModule.cpp
+++ lib/CodeGen/CodeGenModule.cpp
@@ -4900,7 +4900,7 @@
   // Return a bogus pointer if RTTI is disabled, unless it's for EH.
   // FIXME: should we even be calling this method if RTTI is disabled
   // and it's not for EH?
-  if (!ForEH && !getLangOpts().RTTI)
+  if ((!ForEH && !getLangOpts().RTTI) || getLangOpts().CUDAIsDevice)
     return llvm::Constant::getNullValue(Int8PtrTy);
 
   if (ForEH && Ty->isObjCObjectPointerType() &&


-------------- next part --------------
A non-text attachment was scrubbed...
Name: D47694.149989.patch
Type: text/x-patch
Size: 1762 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/cfe-commits/attachments/20180605/4025b6f1/attachment.bin>


More information about the cfe-commits mailing list