Cuda Error:400 Invalid Resource Handle, when passing a stream between functions/classes

Accelerated Computing CUDA CUDA Programming and Performance
1

Greetings,
I am trying to modify an old piece of code heavily using Thrust to work on streams, I have tried to modify the old code but I have ended up with an invalid resource handle error and this means that I am not passing the stream in a correct way and I am not able to spot it, could you guys help me out with it, as I really need another pair of eyes in this case:

(1) I have a class called TriComp,
(2) TriComp has a private member: const cudaStream_t* _stream;
(3) TriComp has no public constructor, but an init(const cudaStream_t* stream) function, init sets _stream

_stream = stream;

(4) In the main the following is done

cudaStream_t testStream;
  cudaStreamCreate(&testStream);
  TriComp triComp;
  triComp.init(&testStream);
  //...
  triComp.compute();

(5)triComp.compute() calls cloudVec.copyFromHost( _input->cV, _stream);
(6)copyFromHost is this in another header file:

void copyFromHost( const thrust::host_vector< T >& inArr, const cudaStream_t* stream)
    // I have only added stream related things
    {
        resize( inArr.size() );
        thrust::copy(
          thrust::cuda::par.on(*stream), //If I comment this one everything works fine, but only on the default stream
          inArr.begin(), inArr.end(), begin() );
        return;
    }

Calling copyFromHost gives Invalid resource handle error, I am sure that I am doing something so very trivial and stupid but I cannot see it :/
Your help will be extremely appreciated. Best Regards,

2

I don’t think its likely that the problem lies in anything you’ve shown. I think it relates to something you haven’t shown or described. For instance if you are doing anything that is multi-GPU oriented, e.g. cudaSetDevice(), that could be relevant to stream usage. This error may occur in a multi-GPU setting where the stream used was not created associated to the current device. I’m not sure it could/would explain what you have shown, however.

3

Greetings,
I am using only a single GPU and choosing device #0.

const int deviceIdx = cutGetMaxGflopsDeviceId();
        CudaSafeCall( cudaSetDevice( deviceIdx ) );
        CudaSafeCall( cudaDeviceReset() );

Unlike the examples in https://devblogs.nvidia.com/cuda-pro-tip-always-set-current-device-avoid-multithreading-bugs/ currently I am not trying anything multithreaded.

If you wish I can share the project as it is an opensource one: https://we.tl/t-UD9Cql7vUC

I have also tried to set device where I try to use streams in desperation:

void copyFromHost( const thrust::host_vector< T >& inArr, const cudaStream_t* stream)
    {
      cudaSetDevice(0);
        resize( inArr.size() );
        thrust::copy(
          thrust::cuda::par.on(*stream),
          inArr.begin(), inArr.end(), begin() );
        return;
    }

Still the same exception, ( nvcuda.dll cuda_runtime.h : 210)

4

cudaDeviceReset wipes out all device context (state). Anything that was allocated or created on the device is gone.

So this won’t work:

void reset()
    {
        Point2HVec().swap( _input.pointVec ); //VD:Host get rid of App class
        SegmentHVec().swap( _input.constraintVec ); //VD:Host
        TriHVec().swap( _output.triVec );
        TriOppHVec().swap( _output.triOppVec );
        cudaDeviceReset();

        return;
    }

    void run()
    {
        // Pick the best CUDA device
        const int deviceIdx = cutGetMaxGflopsDeviceId();
        CudaSafeCall( cudaSetDevice( deviceIdx ) );
        CudaSafeCall( cudaDeviceReset() );
        cudaStream_t testStream;
        cudaStreamCreate(&testStream);  // create a stream


        GpuDel gpuDel;

        for ( int i = 0; i < _runNum; ++i )
        {
            reset();  //implicit destruction of stream
5

Greetings,
Thank you for point that out, a colleague of mine has solved our issue.

  1. device reset is removed
  2. thrust::copy caused the same problems, it is changed with CUDA’s async memcpy, others worked properly
  3. raw pointers assigned to thrust host/device vectors

For now we have a working modification, thank you so much Mr. Crovella
Now all I need to figure out why thrust::copy has failed, but that is for another day

1 Like