<html>
    <head>
      <base href="https://bugs.llvm.org/">
    </head>
    <body><table border="1" cellspacing="0" cellpadding="8">
        <tr>
          <th>Bug ID</th>
          <td><a class="bz_bug_link 
          bz_status_NEW "
   title="NEW - Linking error when calling operator new in CUDA kernel"
   href="https://bugs.llvm.org/show_bug.cgi?id=34360">34360</a>
          </td>
        </tr>

        <tr>
          <th>Summary</th>
          <td>Linking error when calling operator new in CUDA kernel
          </td>
        </tr>

        <tr>
          <th>Product</th>
          <td>clang
          </td>
        </tr>

        <tr>
          <th>Version</th>
          <td>4.0
          </td>
        </tr>

        <tr>
          <th>Hardware</th>
          <td>PC
          </td>
        </tr>

        <tr>
          <th>OS</th>
          <td>MacOS X
          </td>
        </tr>

        <tr>
          <th>Status</th>
          <td>NEW
          </td>
        </tr>

        <tr>
          <th>Severity</th>
          <td>normal
          </td>
        </tr>

        <tr>
          <th>Priority</th>
          <td>P
          </td>
        </tr>

        <tr>
          <th>Component</th>
          <td>CUDA
          </td>
        </tr>

        <tr>
          <th>Assignee</th>
          <td>unassignedclangbugs@nondot.org
          </td>
        </tr>

        <tr>
          <th>Reporter</th>
          <td>ralph_kube@gmx.net
          </td>
        </tr>

        <tr>
          <th>CC</th>
          <td>llvm-bugs@lists.llvm.org
          </td>
        </tr></table>
      <p>
        <div>
        <pre>clang 4.0 fails linking programs which call ::operator new() from CUDA kernels.

This is described in the CUDA documentation here:
<a href="http://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#memory-allocation-and-lifetime">http://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#memory-allocation-and-lifetime</a>

Below is a minimal program that reproduces the error as well as the error
message and the version information.

#include <iostream>
#include <cuda_runtime_api.h>

class myclass
{
public:
   __host__ __device__ myclass(const double _data) : data(_data) {}
   __host__ __device__ ~myclass() 
   {   
       printf("Deleting myclass\n");
   }   
   __host__ __device__ double get_data() const {return(data);}

private:
   const double data;
};



__global__
void init_myclass(myclass** mycs_ptr)
{
   (*mycs_ptr) = new myclass(14.0);
}


__global__
void access_myclass(myclass** mycs_ptr)
{
   printf("I am using data with value = %f\n", (*mycs_ptr) -> get_data());
}

__global__
void delete_myclass(myclass** mycs_ptr)
{
   delete (*mycs_ptr);
}


int main(void)
{
   myclass** myclass_ptr{nullptr};


   init_myclass<<<1, 1>>>(myclass_ptr);

   access_myclass<<<1, 1>>>(myclass_ptr);

   delete_myclass<<<1, 1>>>(myclass_ptr);

   return(0);
}

[1] % clang++ -std=c++14 -o test_new_device test_new_device.cu
-L/Developer/NVIDIA/CUDA-8.0/lib -lcudart
ptxas fatal   : Unresolved extern function '_Znwm'
clang-4.0: error: ptxas command failed with exit code 255 (use -v to see
invocation)

I’m on osx 10.12.5, using
% clang++ --version
clang version 4.0.0 (tags/RELEASE_400/final 297808)
Target: x86_64-apple-darwin16.6.0
Thread model: posix</pre>
        </div>
      </p>


      <hr>
      <span>You are receiving this mail because:</span>

      <ul>
          <li>You are on the CC list for the bug.</li>
      </ul>
    </body>
</html>