From c72bc7cd47ecb5665a287fb88e101f88118f5232 Mon Sep 17 00:00:00 2001 From: Willem Jan Palenstijn Date: Wed, 16 Apr 2014 11:13:40 +0000 Subject: Split up processVol in Vol/Sino cases --- cuda/2d/sart.cu | 14 +++++++------- 1 file changed, 7 insertions(+), 7 deletions(-) (limited to 'cuda/2d/sart.cu') diff --git a/cuda/2d/sart.cu b/cuda/2d/sart.cu index 79c00ef..048661f 100644 --- a/cuda/2d/sart.cu +++ b/cuda/2d/sart.cu @@ -150,14 +150,14 @@ bool SART::precomputeWeights() zeroVolumeData(D_tmpData, tmpPitch, dims); - processVol(D_tmpData, 1.0f, tmpPitch, dims.iVolWidth, dims.iVolHeight); + processVol(D_tmpData, 1.0f, tmpPitch, dims); callFP(D_tmpData, tmpPitch, D_lineWeight, linePitch, 1.0f); cudaFree(D_tmpData); D_tmpData = 0; } - processVol(D_lineWeight, linePitch, dims.iProjDets, dims.iProjAngles); + processSino(D_lineWeight, linePitch, dims); return true; } @@ -185,7 +185,7 @@ bool SART::iterate(unsigned int iterations) // do FP, subtracting projection from sinogram if (useVolumeMask) { cudaMemcpy2D(D_tmpData, sizeof(float)*tmpPitch, D_volumeData, sizeof(float)*volumePitch, sizeof(float)*(dims.iVolWidth), dims.iVolHeight, cudaMemcpyDeviceToDevice); - processVol(D_tmpData, D_maskData, tmpPitch, dims.iVolWidth, dims.iVolHeight); + processVol(D_tmpData, D_maskData, tmpPitch, dims); callFP_SART(D_tmpData, tmpPitch, D_projData, projPitch, angle, -1.0f); } else { callFP_SART(D_volumeData, volumePitch, D_projData, projPitch, angle, -1.0f); @@ -198,15 +198,15 @@ bool SART::iterate(unsigned int iterations) // TODO: Try putting the masking directly in the BP zeroVolumeData(D_tmpData, tmpPitch, dims); callBP_SART(D_tmpData, tmpPitch, D_projData, projPitch, angle); - processVol(D_volumeData, D_maskData, D_tmpData, volumePitch, dims.iVolWidth, dims.iVolHeight); + processVol(D_volumeData, D_maskData, D_tmpData, volumePitch, dims); } else { callBP_SART(D_volumeData, volumePitch, D_projData, projPitch, angle); } if (useMinConstraint) - processVol(D_volumeData, fMinConstraint, volumePitch, dims.iVolWidth, dims.iVolHeight); + processVol(D_volumeData, fMinConstraint, volumePitch, dims); if (useMaxConstraint) - processVol(D_volumeData, fMaxConstraint, volumePitch, dims.iVolWidth, dims.iVolHeight); + processVol(D_volumeData, fMaxConstraint, volumePitch, dims); iteration++; @@ -228,7 +228,7 @@ float SART::computeDiffNorm() // do FP, subtracting projection from sinogram if (useVolumeMask) { cudaMemcpy2D(D_tmpData, sizeof(float)*tmpPitch, D_volumeData, sizeof(float)*volumePitch, sizeof(float)*(dims.iVolWidth), dims.iVolHeight, cudaMemcpyDeviceToDevice); - processVol(D_tmpData, D_maskData, tmpPitch, dims.iVolWidth, dims.iVolHeight); + processVol(D_tmpData, D_maskData, tmpPitch, dims); callFP(D_tmpData, tmpPitch, D_projData, projPitch, -1.0f); } else { callFP(D_volumeData, volumePitch, D_projData, projPitch, -1.0f); -- cgit v1.2.3