diff --git a/BabelBrain/BabelDatasetPreps.py b/BabelBrain/BabelDatasetPreps.py index 0298223..02a70f3 100644 --- a/BabelBrain/BabelDatasetPreps.py +++ b/BabelBrain/BabelDatasetPreps.py @@ -740,11 +740,11 @@ def GetSkullMaskFromSimbNIBSSTL(SimbNIBSDir='4007/4007_keep/m2m_4007_keep/', if VoxelizeFilter is None: while(True): try: - with CodeTimer("skull voxelization",unit='s'): + with CodeTimer("cpu skull voxelization",unit='s'): skull_grid = skull_mesh.voxelized(SpatialStep*0.75,max_iter=30).fill().points.astype(np.float32) - with CodeTimer("brain voxelization",unit='s'): + with CodeTimer("cpu voxelization",unit='s'): csf_grid = csf_mesh.voxelized(SpatialStep*0.75,max_iter=30).fill().points.astype(np.float32) - with CodeTimer("skin voxelization",unit='s'): + with CodeTimer("cpu voxelization",unit='s'): skin_grid = skin_mesh.voxelized(SpatialStep*0.75,max_iter=30).fill().points.astype(np.float32) break except AttributeError as err: @@ -753,11 +753,11 @@ def GetSkullMaskFromSimbNIBSSTL(SimbNIBSDir='4007/4007_keep/m2m_4007_keep/', else: raise err else: - with CodeTimer("cpu skull voxelization",unit='s'): + with CodeTimer("skull voxelization",unit='s'): skull_grid = VoxelizeFilter(skull_mesh,targetResolution=SpatialStep*0.75,GPUBackend=VoxelizeCOMPUTING_BACKEND) - with CodeTimer("cpu brain voxelization",unit='s'): + with CodeTimer("brain voxelization",unit='s'): csf_grid = VoxelizeFilter(csf_mesh,targetResolution=SpatialStep*0.75,GPUBackend=VoxelizeCOMPUTING_BACKEND) - with CodeTimer("cpu skin voxelization",unit='s'): + with CodeTimer("skin voxelization",unit='s'): skin_grid = VoxelizeFilter(skin_mesh,targetResolution=SpatialStep*0.75,GPUBackend=VoxelizeCOMPUTING_BACKEND) diff --git a/BabelBrain/CalculateMaskProcess.py b/BabelBrain/CalculateMaskProcess.py index 96012fa..87fbb40 100644 --- a/BabelBrain/CalculateMaskProcess.py +++ b/BabelBrain/CalculateMaskProcess.py @@ -66,8 +66,7 @@ def __del__(self): BinaryClosing.InitBinaryClosing(DeviceName=devicename,GPUBackend=gpu_backend) DataPreps.InitMedianGPUCallback(MedianFilter.MedianFilter,COMPUTING_BACKEND) - if gpu_backend != 'CUDA': - DataPreps.InitVoxelizeGPUCallback(Voxelize.Voxelize,COMPUTING_BACKEND) + DataPreps.InitVoxelizeGPUCallback(Voxelize.Voxelize,COMPUTING_BACKEND) DataPreps.InitMappingGPUCallback(MappingFilter.MapFilter,COMPUTING_BACKEND) DataPreps.InitResampleGPUCallback(Resample.ResampleFromTo,COMPUTING_BACKEND) DataPreps.InitBinaryClosingGPUCallback(BinaryClosing.BinaryClose,COMPUTING_BACKEND) diff --git a/BabelBrain/GPUFunctions/GPUVoxelize/Voxelize.py b/BabelBrain/GPUFunctions/GPUVoxelize/Voxelize.py index 02e354f..9dcb571 100644 --- a/BabelBrain/GPUFunctions/GPUVoxelize/Voxelize.py +++ b/BabelBrain/GPUFunctions/GPUVoxelize/Voxelize.py @@ -245,9 +245,7 @@ def Voxelize(inputMesh,targetResolution=1333/500e3/6*0.75*1e3,GPUBackend='OpenCL int_params_gpu, np.uint32(gx), np.uint32(gy), - np.uint32(gz), - np.uint32(points_section.size)) - ) + np.uint32(gz))) # Move kernel output data back to host memory points_section=points_section_gpu.get() diff --git a/BabelBrain/GPUFunctions/GPUVoxelize/voxelize.cpp b/BabelBrain/GPUFunctions/GPUVoxelize/voxelize.cpp index 7046903..1311442 100644 --- a/BabelBrain/GPUFunctions/GPUVoxelize/voxelize.cpp +++ b/BabelBrain/GPUFunctions/GPUVoxelize/voxelize.cpp @@ -258,12 +258,9 @@ __kernel void ExtractPoints(__global const unsigned int* voxel_table, const unsigned int * int_params, const unsigned int gx, const unsigned int gy, - const unsigned int gz, - const unsigned int section_size) + const unsigned int gz) { size_t k = (size_t)(blockIdx.x*blockDim.x + threadIdx.x); - - if (k >= section_size) return; #endif #ifdef _METAL