Cuda - 3D块和网格尺寸混淆 - 另一个
Cuda - 3D block & grid dimension confusion - Another one
在下面的简单示例中,我使用cudaMalloc3D
在设备上分配内存,并将我的3D数据的每个体素增加1,这很好,只要我使用对称的3D体积。
主机代码如下所示:
int main(void)
{
typedef float PixelType;
// Set up test data
dim3 image_dimensions = dim3(32, 32, 32);
size_t num_elements = image_dimensions.x * image_dimensions.y * image_dimensions.z;
PixelType *image_data = new float[num_elements];
for(int i = 0; i < num_elements; ++i)
{
image_data[i] = float(i);
}
// Allocate 3D memory on the device
cudaExtent volumeSizeBytes = make_cudaExtent(sizeof(PixelType) * image_dimensions.x, image_dimensions.y, image_dimensions.z);
cudaPitchedPtr devicePitchedPointer;
cudaMalloc3D(&devicePitchedPointer, volumeSizeBytes);
cudaMemset3D(devicePitchedPointer, 1.0f, volumeSizeBytes);
// Copy image data from the host to the device
cudaMemcpy3DParms copy_params_host_to_device = {0};
copy_params_host_to_device.srcPtr = make_cudaPitchedPtr((void *)image_data, sizeof(PixelType) * image_dimensions.x, image_dimensions.y, image_dimensions.z);
copy_params_host_to_device.dstPtr = devicePitchedPointer;
copy_params_host_to_device.extent = volumeSizeBytes;
copy_params_host_to_device.kind = cudaMemcpyHostToDevice;
cudaMemcpy3D(©_params_host_to_device);
// Kernel Launch Configuration
dim3 threads_per_block = dim3(8, 8, 8);
dim3 blocks_per_grid = dim3((image_dimensions.x + threads_per_block.x - 1) / threads_per_block.x, (image_dimensions.y + threads_per_block.y - 1) / threads_per_block.y, (image_dimensions.z + threads_per_block.z - 1) / threads_per_block.z);
extract_patches_from_image_data<<<blocks_per_grid, threads_per_block>>>(devicePitchedPointer, image_dimensions);
cudaDeviceSynchronize();
// Copy image data back from the device to the host
cudaMemcpy3DParms copy_params_device_to_host = {0};
copy_params_device_to_host.srcPtr = devicePitchedPointer;
copy_params_device_to_host.dstPtr = make_cudaPitchedPtr((void *)image_data, sizeof(PixelType) * image_dimensions.x, image_dimensions.y, image_dimensions.z);
copy_params_device_to_host.extent = volumeSizeBytes;
copy_params_device_to_host.kind = cudaMemcpyDeviceToHost;
cudaMemcpy3D(©_params_device_to_host);
// Check image data
for(int i = 0; i < num_elements; ++i)
{
std::cout << "Element: " << i << " - " << image_data[i] << std::endl;
}
// Free Memory
cudaFree(devicePitchedPointer.ptr);
delete [] image_data;
}
所有值递增对应的内核:
__global__ void extract_patches_from_image_data(cudaPitchedPtr devicePitchedPointer, dim3 image_dimensions)
{
// Index Calculation
int x = threadIdx.x + blockDim.x * blockIdx.x;
int y = threadIdx.y + blockDim.y * blockIdx.y;
int z = threadIdx.z + blockDim.z * blockIdx.z;
// Get attributes from device pitched pointer
char *devicePointer = (char *)devicePitchedPointer.ptr;
size_t pitch = devicePitchedPointer.pitch;
size_t slicePitch = pitch * image_dimensions.y;
// Loop over image data
if(z < image_dimensions.z)
{
char *current_slice_index = devicePointer + z * slicePitch;
if(y < image_dimensions.y)
{
// Get data array containing all elements from the current row
PixelType *current_row = (PixelType *)(current_slice_index + y * pitch);
if(x < image_dimensions.x)
{
current_row[x] = current_row[x] + 1.0f;
// Get values of all all neighbors
}
}
}
}
只要我保持我的image_dimensions
对称,例如(32,32,32),一切都很好。当我尝试使用(32,32,33)时,它工作得很好,直到体素33759,以下值保持不变。我现在的问题是我应该如何调整我的代码,以使用非对称数据。
- 我建议做适当的cuda错误检查任何时候你有cuda代码的问题,虽然它不会在这里解决问题。
- 正在传递
float
到cudaMemset3D
。如果你打算将每个浮点数设置为这个值,那就行不通了。cudaMemset3D的工作原理类似主机memset
函数。它接受unsigned char
值并设置unsigned char
数量。您不能使用此方法将float
值正确地初始化为1.0f。但这也不是问题的症结所在。 - 您没有正确使用
make_cudaPitchedPtr
功能。请查看文档。最后两个参数应该分别是x
和y
维度,而不是y
和z
。在你的代码中有两个这样的实例。
通过修改make_cudaPitchedPtr
相关文章:
- 删除一个线程上有数百万个字符串的大型哈希映射会影响另一个线程的性能
- 运行同一解决方案的另一个项目的项目
- 为什么在没有显式默认构造函数的情况下,将另一个结构封装在联合中作为成员的结构不能编译
- 基于另一个成员参数将函数调用从类传递给它的一个成员
- C++从另一个类访问公共静态向量的正确方法是什么
- C++-试图将函数指针推回到另一个CPP文件中的矢量时出错
- 使用std::transform将一个范围的元素添加到另一个范围中
- 输入到文件并输出到另一个文件,并将流文件传递给函数
- 我可以将一个用clang c++11编译的对象与另一个用c++17编译的对象链接起来吗
- 修改函数中的指针(将另一个指针作为参数传递)
- 为什么我不能将一个对象push_back到属于另一个类的对象向量中?
- C++试图读取一个文件并输出到另一个文本文件
- 如何将指针从一个void函数传递到另一个C++
- 如何从另一个文件继承私有成员变量和公共函数
- 使用.find函数在c++中查找字符和另一个字符之间的大小
- 检查 2D 网格的某个元素是否与另一个元素共享对角线、水平线或垂直线
- 如何使用条件计算 3D 网格中从一个点到另一个点的所有路径
- 仅在另一个网眼后面才能渲染网格
- 如何在另一个网格(如蒙版)后面渲染网格
- Cuda - 3D块和网格尺寸混淆 - 另一个