Skip to content

[SYCL][CUDA] PI API CUDA piDeviceGetInfo with Image / Sampler support #1951

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
merged 1 commit into from
Jul 6, 2020
Merged

[SYCL][CUDA] PI API CUDA piDeviceGetInfo with Image / Sampler support #1951

Changes from all commits
Commits
Show all changes
1 commit
Select commit
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
130 changes: 114 additions & 16 deletions sycl/plugins/cuda/pi_cuda.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -904,44 +904,142 @@ pi_result cuda_piDeviceGetInfo(pi_device device, pi_device_info param_name,
}
case PI_DEVICE_INFO_IMAGE_SUPPORT: {
return getInfo(param_value_size, param_value, param_value_size_ret,
PI_FALSE);
PI_TRUE);
}
case PI_DEVICE_INFO_MAX_READ_IMAGE_ARGS: {
return getInfo(param_value_size, param_value, param_value_size_ret, 0);
// This call doesn't match to CUDA as it doesn't have images, but instead
// surfaces and textures. No clear call in the CUDA API to determine this,
// but some searching found as of SM 2.x 128 are supported.
return getInfo(param_value_size, param_value, param_value_size_ret, 128u);
}
case PI_DEVICE_INFO_MAX_WRITE_IMAGE_ARGS: {
return getInfo(param_value_size, param_value, param_value_size_ret, 0u);
// This call doesn't match to CUDA as it doesn't have images, but instead
// surfaces and textures. No clear call in the CUDA API to determine this,
// but some searching found as of SM 2.x 128 are supported.
return getInfo(param_value_size, param_value, param_value_size_ret, 128u);
}
case PI_DEVICE_INFO_IMAGE2D_MAX_HEIGHT: {
return getInfo(param_value_size, param_value, param_value_size_ret,
size_t(0));
// Take the smaller of maximum surface and maximum texture height.
int tex_height = 0;
cl::sycl::detail::pi::assertion(
cuDeviceGetAttribute(&tex_height,
CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE2D_HEIGHT,
device->get()) == CUDA_SUCCESS);
cl::sycl::detail::pi::assertion(tex_height >= 0);
int surf_height = 0;
cl::sycl::detail::pi::assertion(
cuDeviceGetAttribute(&surf_height,
CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACE2D_HEIGHT,
device->get()) == CUDA_SUCCESS);
cl::sycl::detail::pi::assertion(surf_height >= 0);

int min = std::min(tex_height, surf_height);

return getInfo(param_value_size, param_value, param_value_size_ret, min);
}
case PI_DEVICE_INFO_IMAGE2D_MAX_WIDTH: {
return getInfo(param_value_size, param_value, param_value_size_ret,
size_t(0));
// Take the smaller of maximum surface and maximum texture width.
int tex_width = 0;
cl::sycl::detail::pi::assertion(
cuDeviceGetAttribute(&tex_width,
CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE2D_WIDTH,
device->get()) == CUDA_SUCCESS);
cl::sycl::detail::pi::assertion(tex_width >= 0);
int surf_width = 0;
cl::sycl::detail::pi::assertion(
cuDeviceGetAttribute(&surf_width,
CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACE2D_WIDTH,
device->get()) == CUDA_SUCCESS);
cl::sycl::detail::pi::assertion(surf_width >= 0);

int min = std::min(tex_width, surf_width);

return getInfo(param_value_size, param_value, param_value_size_ret, min);
}
case PI_DEVICE_INFO_IMAGE3D_MAX_HEIGHT: {
return getInfo(param_value_size, param_value, param_value_size_ret,
size_t(0));
// Take the smaller of maximum surface and maximum texture height.
int tex_height = 0;
cl::sycl::detail::pi::assertion(
cuDeviceGetAttribute(&tex_height,
CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE3D_HEIGHT,
device->get()) == CUDA_SUCCESS);
cl::sycl::detail::pi::assertion(tex_height >= 0);
int surf_height = 0;
cl::sycl::detail::pi::assertion(
cuDeviceGetAttribute(&surf_height,
CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACE3D_HEIGHT,
device->get()) == CUDA_SUCCESS);
cl::sycl::detail::pi::assertion(surf_height >= 0);

int min = std::min(tex_height, surf_height);

return getInfo(param_value_size, param_value, param_value_size_ret, min);
}
case PI_DEVICE_INFO_IMAGE3D_MAX_WIDTH: {
return getInfo(param_value_size, param_value, param_value_size_ret,
size_t(0));
// Take the smaller of maximum surface and maximum texture width.
int tex_width = 0;
cl::sycl::detail::pi::assertion(
cuDeviceGetAttribute(&tex_width,
CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE3D_WIDTH,
device->get()) == CUDA_SUCCESS);
cl::sycl::detail::pi::assertion(tex_width >= 0);
int surf_width = 0;
cl::sycl::detail::pi::assertion(
cuDeviceGetAttribute(&surf_width,
CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACE3D_WIDTH,
device->get()) == CUDA_SUCCESS);
cl::sycl::detail::pi::assertion(surf_width >= 0);

int min = std::min(tex_width, surf_width);

return getInfo(param_value_size, param_value, param_value_size_ret, min);
}
case PI_DEVICE_INFO_IMAGE3D_MAX_DEPTH: {
return getInfo(param_value_size, param_value, param_value_size_ret,
size_t(0));
// Take the smaller of maximum surface and maximum texture depth.
int tex_depth = 0;
cl::sycl::detail::pi::assertion(
cuDeviceGetAttribute(&tex_depth,
CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE3D_DEPTH,
device->get()) == CUDA_SUCCESS);
cl::sycl::detail::pi::assertion(tex_depth >= 0);
int surf_depth = 0;
cl::sycl::detail::pi::assertion(
cuDeviceGetAttribute(&surf_depth,
CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACE3D_DEPTH,
device->get()) == CUDA_SUCCESS);
cl::sycl::detail::pi::assertion(surf_depth >= 0);

int min = std::min(tex_depth, surf_depth);

return getInfo(param_value_size, param_value, param_value_size_ret, min);
}
case PI_DEVICE_INFO_IMAGE_MAX_BUFFER_SIZE: {
return getInfo(param_value_size, param_value, param_value_size_ret,
size_t(0));
// Take the smaller of maximum surface and maximum texture width.
int tex_width = 0;
cl::sycl::detail::pi::assertion(
cuDeviceGetAttribute(&tex_width,
CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE1D_WIDTH,
device->get()) == CUDA_SUCCESS);
cl::sycl::detail::pi::assertion(tex_width >= 0);
int surf_width = 0;
cl::sycl::detail::pi::assertion(
cuDeviceGetAttribute(&surf_width,
CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACE1D_WIDTH,
device->get()) == CUDA_SUCCESS);
cl::sycl::detail::pi::assertion(surf_width >= 0);

int min = std::min(tex_width, surf_width);

return getInfo(param_value_size, param_value, param_value_size_ret, min);
}
case PI_DEVICE_INFO_IMAGE_MAX_ARRAY_SIZE: {
return getInfo(param_value_size, param_value, param_value_size_ret,
size_t(0));
}
case PI_DEVICE_INFO_MAX_SAMPLERS: {
return getInfo(param_value_size, param_value, param_value_size_ret, 0u);
// This call is kind of meaningless for cuda, as samplers don't exist.
// Closest thing is textures, which is 128.
return getInfo(param_value_size, param_value, param_value_size_ret, 128u);
}
case PI_DEVICE_INFO_MAX_PARAMETER_SIZE: {
// https://docs.nvidia.com/cuda/cuda-c-programming-guide/#function-parameters
Expand Down