title: cuda streams: callback not getting called after stream execution
body: <p>I am having trouble with the cuda streams callback function. The problem is that it does not execute after the kernel is run. I am encapsulating the cuda streams in a C++ class as follows:</p>
  
  <pre><code>class Stream
  {
  public:
      Stream();
      void run(float *input, int points);
      ~Stream();
      static void CUDART_CB callback(cudaStream_t stream,
                                     cudaError_t status,
                                     void * user_data);
  
  private:
      void callback_function();
  
      cudaStream_t s;
  
  };
  </code></pre>
  
  <p>Now, I have the callback function setup as suggested before as follows:</p>
  
  <pre><code>void CUDART_CB StreamWorkflow::callback(cudaStream_t stream,
                                      cudaError_t status,
                                      void * user_data)
  {
      Stream* thiz = (Stream *)(user_data);
      thiz-&gt;callback_function();
  }
  </code></pre>
  
  <p>And the constructor sets up the stream and attaches the callback</p>
  
  <pre><code>Stream::Stream()
  {
      checkCudaErrors(cudaStreamCreate(&amp;s));
      checkCudaErrors(cudaStreamAddCallback(s, Stream::callback, this, 0));
  }
  </code></pre>
  
  <p>Now the run method calls the cuda kernel on this stream and I know this executes fine. It is implemented like:</p>
  
  <pre><code>void Stream::run(float *f, int p)
  {
      dim3 block(16, 16);
      dim3 grid((int)ceil(double(p) / 256.0));
      my_kernel(f, p, grid, block, s);
      checkCudaErrors(cudaDeviceSynchronize());
      getLastCudaError("kernel launch failed");
  }
  </code></pre>
  
  <p>My problem is that the callback only gets called when the constructor gets executed. So, as soon as the object is created the callback_function() gets executed through the registered callback(). The kernel gets executed multiple times by the host and it never executes the callback function when it is finished. I can see the kernel is executed successfully but the callback never comes. </p>
  
  <p>Do I have to register the callback <strong>everytime</strong> the kernel is run (so register this callback before executing the run function) or am I doing something really wrong here?</p>