diff options
| author | Willem Jan Palenstijn <Willem.Jan.Palenstijn@cwi.nl> | 2021-11-16 13:44:13 +0100 |
|---|---|---|
| committer | Willem Jan Palenstijn <Willem.Jan.Palenstijn@cwi.nl> | 2021-11-16 14:06:30 +0100 |
| commit | 39582115bc93b5435d25e56891815ae7cb1898fd (patch) | |
| tree | 6c728e1125961fc04ba6f77bf9af637925825f5b /cuda/2d/util.cu | |
| parent | b492e3d049e300132d2f22eee7922ff308342a84 (diff) | |
| download | astra-39582115bc93b5435d25e56891815ae7cb1898fd.tar.gz astra-39582115bc93b5435d25e56891815ae7cb1898fd.tar.bz2 astra-39582115bc93b5435d25e56891815ae7cb1898fd.tar.xz astra-39582115bc93b5435d25e56891815ae7cb1898fd.zip | |
Remove cudaTextForceKernelsCompletion
Diffstat (limited to 'cuda/2d/util.cu')
| -rw-r--r-- | cuda/2d/util.cu | 17 |
1 files changed, 2 insertions, 15 deletions
diff --git a/cuda/2d/util.cu b/cuda/2d/util.cu index a75e5ab..ac360f0 100644 --- a/cuda/2d/util.cu +++ b/cuda/2d/util.cu @@ -216,7 +216,7 @@ float dotProduct2D(float* D_data, unsigned int pitch, // Step 1: reduce 2D from image to a single vector, taking sum of squares reduce2D<<< dimGrid2, dimBlock2, shared_mem2>>>(D_data, D_buf, pitch, width, height); - cudaTextForceKernelsCompletion(); + checkCuda(cudaThreadSynchronize(), "dotProduct2D reduce2D"); // Step 2: reduce 1D: add up elements in vector if (bx * by > 512) @@ -233,26 +233,13 @@ float dotProduct2D(float* D_data, unsigned int pitch, float x; cudaMemcpy(&x, D_res, 4, cudaMemcpyDeviceToHost); - cudaTextForceKernelsCompletion(); + checkCuda(cudaThreadSynchronize(), "dotProduct2D"); cudaFree(D_buf); return x; } - -bool cudaTextForceKernelsCompletion() -{ - cudaError_t returnedCudaError = cudaThreadSynchronize(); - - if(returnedCudaError != cudaSuccess) { - ASTRA_ERROR("Failed to force completion of cuda kernels: %d: %s.", returnedCudaError, cudaGetErrorString(returnedCudaError)); - return false; - } - - return true; -} - bool checkCuda(cudaError_t err, const char *msg) { if (err != cudaSuccess) { |