summaryrefslogtreecommitdiffstats
path: root/cuda/2d/em.cu
diff options
context:
space:
mode:
authorWillem Jan Palenstijn <WillemJan.Palenstijn@uantwerpen.be>2014-04-16 11:13:46 +0000
committerwpalenst <WillemJan.Palenstijn@uantwerpen.be>2014-04-16 11:13:46 +0000
commite0b3ad8e57f269e34085ba319aa399ee3476811a (patch)
treee34426606814f3df61f4e590c76d17e808925a37 /cuda/2d/em.cu
parentc72bc7cd47ecb5665a287fb88e101f88118f5232 (diff)
downloadastra-e0b3ad8e57f269e34085ba319aa399ee3476811a.tar.gz
astra-e0b3ad8e57f269e34085ba319aa399ee3476811a.tar.bz2
astra-e0b3ad8e57f269e34085ba319aa399ee3476811a.tar.xz
astra-e0b3ad8e57f269e34085ba319aa399ee3476811a.zip
Replace direct cudaMemcpy2D calls by utility functions
Diffstat (limited to 'cuda/2d/em.cu')
-rw-r--r--cuda/2d/em.cu4
1 files changed, 2 insertions, 2 deletions
diff --git a/cuda/2d/em.cu b/cuda/2d/em.cu
index c75f250..ebb76b5 100644
--- a/cuda/2d/em.cu
+++ b/cuda/2d/em.cu
@@ -150,11 +150,11 @@ bool EM::iterate(unsigned int iterations)
float EM::computeDiffNorm()
{
// copy sinogram to projection data
- cudaMemcpy2D(D_projData, sizeof(float)*projPitch, D_sinoData, sizeof(float)*sinoPitch, sizeof(float)*(dims.iProjDets), dims.iProjAngles, cudaMemcpyDeviceToDevice);
+ duplicateProjectionData(D_projData, D_sinoData, sinoPitch, dims);
// 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);
+ duplicateVolumeData(D_tmpData, D_volumeData, volumePitch, dims);
processVol<opMul>(D_tmpData, D_maskData, tmpPitch, dims);
callFP(D_tmpData, tmpPitch, D_projData, projPitch, -1.0f);
} else {