From 32ad08f8d7f961b97588c4573f5474b4fa8debdd Mon Sep 17 00:00:00 2001
From: kradchen <chenjuijun@realcan.cn>
Date: Tue, 26 Dec 2023 15:20:26 +0800
Subject: [PATCH] Add cudaDeviceSynchronize to detection

---
 src/transmissionReconstruction/detection/detection.cu | 5 ++++-
 1 file changed, 4 insertions(+), 1 deletion(-)

diff --git a/src/transmissionReconstruction/detection/detection.cu b/src/transmissionReconstruction/detection/detection.cu
index 63502c8..b82501c 100644
--- a/src/transmissionReconstruction/detection/detection.cu
+++ b/src/transmissionReconstruction/detection/detection.cu
@@ -55,7 +55,7 @@ CudaMatrix Recon::calculateAttenuationCuda(const CudaMatrix &ascans,
                                                     ascans.getData(), ascans2.getData(),
                                                     ascansRef.getData(), ascansRef2.getData(),
                                                     ascans2.getDimSize(0));
-
+    cudaDeviceSynchronize();
     auto pulseEnergy = Aurora::sum(ascans2^2);
     auto pulseEnergyEmpty = Aurora::sum(ascansRef2^2);
 
@@ -86,6 +86,7 @@ Recon::detectAttVectorizedCuda(const CudaMatrix &Ascan, const CudaMatrix &AscanR
                                 endPos.getData(), startPosRef.getData(), endPosRef.getData(),
                                 tof.getData(), tof2.getData(), sizeAscan.getData(), sampleRate,
                                 offsetElectronicSamples,detectionWindowATT);
+    cudaDeviceSynchronize();
     return calculateAttenuationCuda(envelopeOfAScan, startPos, endPos,
                                 envelopeOfReferenceAScan, startPosRef,
                                 endPosRef);
@@ -189,6 +190,7 @@ Recon::TimeWindowResultC Recon::applyTimeWindowing(const Aurora::CudaMatrix &Asc
                             calcResult.endSearch.getData(),AscanBlock.getData(),AscanBlockProcessed.getData(),
                             AscanBlock.getDimSize(0));
         }
+        cudaDeviceSynchronize();
         Recon::TimeWindowResultC result;
         result.startSearch = calcResult.startSearch;
         result.AscanBlockProcessed = AscanBlockProcessed;
@@ -306,6 +308,7 @@ Recon::DetectResultC Recon::detectTofAndAtt(
         }
         auto shiftInSamples = zerosCuda(1, c1.getDimSize(1));
         findMaxIndexKernel<<<c.getDimSize(1),256>>>(c.getData(),c.getDimSize(0),shiftInSamples.getData(),maxlag);
+        cudaDeviceSynchronize();
         if (useTimeWindowing) {
             shiftInSamples = shiftInSamples - diffStartSearch;
         }