<table border="1" cellspacing="0" cellpadding="8">
<tr>
<th>Issue</th>
<td>
<a href=https://github.com/llvm/llvm-project/issues/114048>114048</a>
</td>
</tr>
<tr>
<th>Summary</th>
<td>
Error when compiling CUDA code with libc++ and C++ standard 26
</td>
</tr>
<tr>
<th>Labels</th>
<td>
libc++
</td>
</tr>
<tr>
<th>Assignees</th>
<td>
</td>
</tr>
<tr>
<th>Reporter</th>
<td>
OgnianM
</td>
</tr>
</table>
<pre>
The error:
```
/usr/local/lib/clang/20/include/cuda_wrappers/new:94:25: error: __device__ function 'operator new' cannot overload __host__ __device__ function 'operator new'
94 | __device__ inline void *operator new(__SIZE_TYPE__, void *__ptr) CUDA_NOEXCEPT {
| ^
/usr/local/bin/../include/c++/v1/new:252:1: note: previous declaration is here
252 | operator new(std::size_t, void* __p) _NOEXCEPT {
| ^
```
Looks like it's caused by the `_LIBCPP_CONSTEXPR_SINCE_CXX26` macro making `operator new(std::size_t, void* __p)` implicitly a __host__ __device__ function when CXX >= 26.
The simplest fix would be wrapping the device overrides for placement new in `cuda_wrappers/new` in an `#if` like so
```
// Device overrides for placement new and delete.
#if _LIBCPP_STD_VER < 26
__device__ inline void *operator new(__SIZE_TYPE__, void *__ptr) CUDA_NOEXCEPT {
return __ptr;
}
__device__ inline void *operator new[](__SIZE_TYPE__, void *__ptr) CUDA_NOEXCEPT {
return __ptr;
}
#endif
```
</pre>
<img width="1px" height="1px" alt="" src="http://email.email.llvm.org/o/eJy8VE1v4zYQ_TX0ZbCGNLKs9UEHR5aBANsk2KSF2wtBkWOLDU0KJOU0_fUF5cSbTb92Dy1AiBLFmXlv5s2IEPTBEtWsvGLlZibG2Dtf3x6sFvaHWefUc_3QE5D3zrNizbINy9Zsmb2s8ydux-AZbo2TwqRddwy30gh7YLjFjOFWW2lGRel4VII_eTEM5APDraUnVqxXC1assWTF-hILOFd00pI4h_1oZdTOAsPKDeRFdB6SJVYghbUugjuRN04o4Lx3IXL-jfZnDgCwWgCrmrdW2hptCU5OK2C4_trwI-f317-0_OHnu5Zzhs3lHudD9AxX0Py4WfOb23bXtHcPwKqrSyyAKRYr279JYactw-18_nXqGF5Na3vKL5nDElmxzlPCrIuU9sHTSbsxgCJphBcTcx2gJ0-vELDECcI7UiGqVOZiHfTvxOMrLYapHEMi9Y2E3klken5y7jGA0Y8EOjKsAkgxBlLQPUPsCdgy45-ur5q7O97c3tw_tLu7z_z--qZpebPb4ZItMzgK6R0cxaO2h2TwvfiTD30cjJY6mmcQ_yyXp54sNLsdsKJlxQZwOX9LKLVGSN4oRNjr3-DJjUZBRzAJPEFMvM5-J4V6rSjA3nkYjJB0JBsTctA2kfmr3kh4LYjpP8NC79PJlMPg3mL5c1My3MLm30MLq0CRoUiv3FIUeK3E_cOG_9R-BlY0gMvzjf-jRzzF0Vs4Xyxejlm1-S4E01T7L3EwLMgqvX9Xg5mqC7UqVmJGdV4VGWZV-TGb9bVcSkFYlaUUy6zKi2pPnZSU475arQpZzHSNGS7yDFd5kWO5nFOpSpHle5EJ6vIO2SKjo9BmbszpOHf-MNMhjFTn-SJbfJwZ0ZEJ0zxHNLq7zAxM893XyepDNx4CW2RGhxi--Ik6GqrbNH7PwpfuOGiTVJyyA9Ipgicde_jid1JP8_IeorBKeAW4nI3e1H2MQ0jtOGnxoGM_dnPpjmnOmdPr9mHw7leSMY26xCQJ_4XMqcY_AgAA__9npQV7">