From 27a1ff287fe3b8caac676d51e968a044cf76cfd1 Mon Sep 17 00:00:00 2001
From: qon <qon@jwdt.org>
Date: Sun, 6 Apr 2014 22:24:31 +0200
Subject: [PATCH] fix asynchronous GPU processing, synchronization was one step
too late before
---
bioem_cuda.cu | 10 ++++++++--
1 file changed, 8 insertions(+), 2 deletions(-)
diff --git a/bioem_cuda.cu b/bioem_cuda.cu
index e890c62..c1c4230 100644
--- a/bioem_cuda.cu
+++ b/bioem_cuda.cu
@@ -84,7 +84,6 @@ int bioem_cuda::compareRefMaps(int iProjectionOut, int iConv, const bioem_map& c
if (GPUAsync)
{
checkCudaErrors(cudaEventSynchronize(cudaEvent[iConv & 1]));
- checkCudaErrors(cudaEventRecord(cudaEvent[iConv & 1], cudaStream));
}
checkCudaErrors(cudaMemcpyAsync(pConvMap_device[iConv & 1], &conv_map, sizeof(bioem_map), cudaMemcpyHostToDevice, cudaStream));
@@ -133,7 +132,14 @@ int bioem_cuda::compareRefMaps(int iProjectionOut, int iConv, const bioem_map& c
cout << "Invalid GPU Algorithm selected\n";
exit(1);
}
- if (GPUAsync == 0) checkCudaErrors(cudaStreamSynchronize(cudaStream));
+ if (GPUAsync)
+ {
+ checkCudaErrors(cudaEventRecord(cudaEvent[iConv & 1], cudaStream));
+ }
+ else
+ {
+ checkCudaErrors(cudaStreamSynchronize(cudaStream));
+ }
return(0);
}
--
GitLab