[PATCH] D97708: [CUDA] Remove `noreturn` attribute from __assertfail().

[Artem Belevich via Phabricator via cfe-commits]

This revision was landed with ongoing or failed builds.
This revision was automatically updated to reflect the committed changes.
Closed by commit rG32e064527623: [CUDA] Remove `noreturn` attribute from __assertfail(). (authored by tra).

Repository:
  rG LLVM Github Monorepo

CHANGES SINCE LAST ACTION
  https://reviews.llvm.org/D97708/new/

https://reviews.llvm.org/D97708

Files:
  clang/lib/Headers/__clang_cuda_runtime_wrapper.h


Index: clang/lib/Headers/__clang_cuda_runtime_wrapper.h
===================================================================
--- clang/lib/Headers/__clang_cuda_runtime_wrapper.h
+++ clang/lib/Headers/__clang_cuda_runtime_wrapper.h
@@ -349,9 +349,14 @@
 __device__ int vprintf(const char *, const char *);
 __device__ void free(void *) __attribute((nothrow));
 __device__ void *malloc(size_t) __attribute((nothrow)) __attribute__((malloc));
+
+// __assertfail() used to have a `noreturn` attribute. Unfortunately that
+// contributed to triggering the longstanding bug in ptxas when assert was used
+// in sufficiently convoluted code. See
+// https://bugs.llvm.org/show_bug.cgi?id=27738 for the details.
 __device__ void __assertfail(const char *__message, const char *__file,
                              unsigned __line, const char *__function,
-                             size_t __charSize) __attribute__((noreturn));
+                             size_t __charSize);
 
 // In order for standard assert() macro on linux to work we need to
 // provide device-side __assert_fail()


-------------- next part --------------
A non-text attachment was scrubbed...
Name: D97708.327285.patch
Type: text/x-patch
Size: 1084 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/cfe-commits/attachments/20210301/1608bd50/attachment.bin>