Gaudi::CUDA::Detail Namespace Reference

Functions
void *	allocateWithStream (std::size_t size, Stream &stream)
void *	allocateNoStream (std::size_t size, Gaudi::AsynchronousAlgorithm *parent)
void	freeWithStream (void *ptr, Stream &stream)
void	freeNoStream (void *ptr)
void	copyHostToDeviceWithStream (void *devPtr, const void *hstPtr, std::size_t size, Stream &stream)
void	copyHostToDeviceNoStream (void *devPtr, const void *hstPtr, std::size_t size)
void	copyDeviceToHostWithStream (void *hstPtr, const void *devPtr, std::size_t size, Stream &stream)
void	copyDeviceToHostNoStream (void *hstPtr, const void *devPtr, std::size_t size)
void	copyDeviceToDeviceWithStream (void *destDevPtr, const void *srcDevPtr, std::size_t size, Stream &stream)
void	copyDeviceToDeviceNoStream (void *destDevPtr, const void *srcDevPtr, std::size_t size)

Function Documentation

allocateNoStream()

void * Gaudi::CUDA::Detail::allocateNoStream	(	std::size_t	size,
		Gaudi::AsynchronousAlgorithm *	parent )

Definition at line 72 of file CUDADeviceArray.cpp.

                                                                               {
    void*       devPtr = nullptr;
    cudaError_t err    = cudaSuccess;
    do {
      err = cudaMalloc( &devPtr, size );
      if ( err == cudaSuccess ) { break; }
      if ( err == cudaErrorMemoryAllocation ) {
        cudaGetLastError();
        // If called from an AsynchronousAlgorithm, wait as in the with stream variant
        // Otherwise, the thread should sleep
        if ( parent != nullptr ) {
          parent
              ->decorateSuspension( [&]() {
                std::unique_lock lck( gpu_mem_mtx );
                gpu_mem_cv.wait( lck );
                return StatusCode::SUCCESS;
              } )
              .orThrow( "Error supending while waiting for memory", DEVARREXC );
        } else {
          std::this_thread::sleep_for( 100ms );
        }
      } else {
        throw GaudiException( err_fmt( err, __FILE__, __LINE__ ), DEVARREXC, StatusCode::FAILURE );
      }
    } while ( err == cudaErrorMemoryAllocation );
    return devPtr;
  }

allocateWithStream()

void * Gaudi::CUDA::Detail::allocateWithStream	(	std::size_t	size,
		Stream &	stream )

Definition at line 48 of file CUDADeviceArray.cpp.

                                                             {
    const Gaudi::AsynchronousAlgorithm* parent = stream.asyncParent();
    void*                               devPtr = nullptr;
    cudaError_t                         err    = cudaSuccess;
    do {
      err = cudaMallocAsync( &devPtr, size, stream );
      if ( err == cudaSuccess ) { break; }
      if ( err == cudaErrorMemoryAllocation && parent != nullptr ) {
        // Waiting and retrying only works if we're in an asynchronous algorithm
        cudaGetLastError();
        parent
            ->decorateSuspension( [&]() {
              std::unique_lock lck( gpu_mem_mtx );
              gpu_mem_cv.wait( lck );
              return StatusCode::SUCCESS;
            } )
            .orThrow( "Error suspending while waiting for memory", DEVARREXC );
      } else {
        throw GaudiException( err_fmt( err, __FILE__, __LINE__ ), DEVARREXC, StatusCode::FAILURE );
      }
    } while ( err == cudaErrorMemoryAllocation );
    return devPtr;
  }

copyDeviceToDeviceNoStream()

void Gaudi::CUDA::Detail::copyDeviceToDeviceNoStream	(	void *	destDevPtr,
		const void *	srcDevPtr,
		std::size_t	size )

Definition at line 155 of file CUDADeviceArray.cpp.

                                                                                             {
    cudaError_t err = cudaMemcpy( destDevPtr, srcDevPtr, size, cudaMemcpyDeviceToDevice );
    if ( err != cudaSuccess ) {
      throw GaudiException( err_fmt( err, __FILE__, __LINE__ ), DEVARREXC, StatusCode::FAILURE );
    }
  }

copyDeviceToDeviceWithStream()

void Gaudi::CUDA::Detail::copyDeviceToDeviceWithStream	(	void *	destDevPtr,
		const void *	srcDevPtr,
		std::size_t	size,
		Stream &	stream )

Definition at line 148 of file CUDADeviceArray.cpp.

                                                                                                               {
    cudaError_t err = cudaMemcpyAsync( destDevPtr, srcDevPtr, size, cudaMemcpyDeviceToDevice, stream );
    if ( err != cudaSuccess ) {
      throw GaudiException( err_fmt( err, __FILE__, __LINE__ ), DEVARREXC, StatusCode::FAILURE );
    }
  }

copyDeviceToHostNoStream()

void Gaudi::CUDA::Detail::copyDeviceToHostNoStream	(	void *	hstPtr,
		const void *	devPtr,
		std::size_t	size )

Definition at line 141 of file CUDADeviceArray.cpp.

                                                                                    {
    cudaError_t err = cudaMemcpy( hstPtr, devPtr, size, cudaMemcpyDeviceToHost );
    if ( err != cudaSuccess ) {
      throw GaudiException( err_fmt( err, __FILE__, __LINE__ ), DEVARREXC, StatusCode::FAILURE );
    }
  }

copyDeviceToHostWithStream()

void Gaudi::CUDA::Detail::copyDeviceToHostWithStream	(	void *	hstPtr,
		const void *	devPtr,
		std::size_t	size,
		Stream &	stream )

Definition at line 132 of file CUDADeviceArray.cpp.

                                                                                                      {
    cudaError_t err = cudaMemcpyAsync( hstPtr, devPtr, size, cudaMemcpyDeviceToHost, stream );
    if ( err != cudaSuccess ) {
      throw GaudiException( err_fmt( err, __FILE__, __LINE__ ), DEVARREXC, StatusCode::FAILURE );
    }
    // await stream to avoid deleting host memory before copy is done
    stream.await().orThrow( "Await error", DEVARREXC );
  }

copyHostToDeviceNoStream()

void Gaudi::CUDA::Detail::copyHostToDeviceNoStream	(	void *	devPtr,
		const void *	hstPtr,
		std::size_t	size )

Definition at line 125 of file CUDADeviceArray.cpp.

                                                                                    {
    cudaError_t err = cudaMemcpy( devPtr, hstPtr, size, cudaMemcpyHostToDevice );
    if ( err != cudaSuccess ) {
      throw GaudiException( err_fmt( err, __FILE__, __LINE__ ), DEVARREXC, StatusCode::FAILURE );
    }
  }

copyHostToDeviceWithStream()

void Gaudi::CUDA::Detail::copyHostToDeviceWithStream	(	void *	devPtr,
		const void *	hstPtr,
		std::size_t	size,
		Stream &	stream )

Definition at line 116 of file CUDADeviceArray.cpp.

                                                                                                      {
    cudaError_t err = cudaMemcpyAsync( devPtr, hstPtr, size, cudaMemcpyHostToDevice, stream );
    if ( err != cudaSuccess ) {
      throw GaudiException( err_fmt( err, __FILE__, __LINE__ ), DEVARREXC, StatusCode::FAILURE );
    }
    // await stream to avoid deleting host memory before copy is done
    stream.await().orThrow( "Await error", DEVARREXC );
  }

freeNoStream()

void Gaudi::CUDA::Detail::freeNoStream

(

void *

ptr

)

Definition at line 108 of file CUDADeviceArray.cpp.

                                 {
    cudaError_t err = cudaFree( ptr );
    if ( err != cudaSuccess ) {
      throw GaudiException( err_fmt( err, __FILE__, __LINE__ ), DEVARREXC, StatusCode::FAILURE );
    }
    gpu_mem_cv.notify_all();
  }

freeWithStream()

void Gaudi::CUDA::Detail::freeWithStream	(	void *	ptr,
		Stream &	stream )

Definition at line 100 of file CUDADeviceArray.cpp.

                                                   {
    cudaError_t err = cudaFreeAsync( ptr, stream );
    if ( err != cudaSuccess ) {
      throw GaudiException( err_fmt( err, __FILE__, __LINE__ ), DEVARREXC, StatusCode::FAILURE );
    }
    gpu_mem_cv.notify_all();
  }