From e17423a24e2502f7b956b3975da08aa87377eb0d Mon Sep 17 00:00:00 2001 From: ROCm CI Service Account <66695075+rocm-ci@users.noreply.github.com> Date: Mon, 9 May 2022 21:16:20 +0530 Subject: [PATCH] SWDEV-305992 - Improve texture tests (#2646) Enable more texture supports for hip-rocclr. Skip texture tests on device that doesn't support images. SWDEV-322257 - Fix issues of image tests skipped on NVidia devices. Change-Id: Ia99d06b1e97fc945f1b740e47710f4dcd70f38ca [ROCm/hip-tests commit: 70ccfa39adb9f820c1e57e070424cdf515cb73d0] --- .../hip-tests/catch/include/hip_test_common.hh | 18 ++++++++++++++++++ .../hipCreateTextureObject_ArgValidation.cc | 2 ++ .../texture/hipCreateTextureObject_Array.cc | 4 ++++ .../texture/hipCreateTextureObject_Linear.cc | 2 ++ .../texture/hipCreateTextureObject_Pitch2D.cc | 2 ++ .../unit/texture/hipTextureObjFetchVector.cc | 4 ++++ .../11_texture_driver/tex2dKernel.cpp | 2 ++ .../11_texture_driver/texture2dDrv.cpp | 13 +++++++++++++ 8 files changed, 47 insertions(+) diff --git a/projects/hip-tests/catch/include/hip_test_common.hh b/projects/hip-tests/catch/include/hip_test_common.hh index 68817ef1c2..e767f07387 100644 --- a/projects/hip-tests/catch/include/hip_test_common.hh +++ b/projects/hip-tests/catch/include/hip_test_common.hh @@ -115,4 +115,22 @@ static inline int RAND_R(unsigned* rand_seed) return rand_r(rand_seed); #endif } + +inline bool isImageSupported() { + int imageSupport = 1; +#ifdef __HIP_PLATFORM_AMD__ + int device; + HIP_CHECK(hipGetDevice(&device)); + HIPCHECK(hipDeviceGetAttribute(&imageSupport, hipDeviceAttributeImageSupport, + device)); +#endif + return imageSupport != 0; } + +} + +// This must be called in the beginning of image test app's main() to indicate whether image +// is supported. +#define checkImageSupport() \ + if (!HipTest::isImageSupported()) \ + { printf("Texture is not support on the device. Skipped.\n"); return; } diff --git a/projects/hip-tests/catch/unit/texture/hipCreateTextureObject_ArgValidation.cc b/projects/hip-tests/catch/unit/texture/hipCreateTextureObject_ArgValidation.cc index 91a0827127..c7570aea03 100644 --- a/projects/hip-tests/catch/unit/texture/hipCreateTextureObject_ArgValidation.cc +++ b/projects/hip-tests/catch/unit/texture/hipCreateTextureObject_ArgValidation.cc @@ -25,6 +25,8 @@ THE SOFTWARE. * Validate argument list of texture object api. */ TEST_CASE("Unit_hipCreateTextureObject_ArgValidation") { + checkImageSupport(); + float *texBuf; hipError_t ret; constexpr int xsize = 32; diff --git a/projects/hip-tests/catch/unit/texture/hipCreateTextureObject_Array.cc b/projects/hip-tests/catch/unit/texture/hipCreateTextureObject_Array.cc index 037efea1af..da42c7878e 100644 --- a/projects/hip-tests/catch/unit/texture/hipCreateTextureObject_Array.cc +++ b/projects/hip-tests/catch/unit/texture/hipCreateTextureObject_Array.cc @@ -23,6 +23,8 @@ THE SOFTWARE. * Validates Array Resource texture object with negative/functional tests. */ TEST_CASE("Unit_hipCreateTextureObject_ArrayResource") { + checkImageSupport(); + hipError_t ret; hipResourceDesc resDesc; hipTextureDesc texDesc; @@ -47,6 +49,8 @@ TEST_CASE("Unit_hipCreateTextureObject_ArrayResource") { * with negative/functional tests. */ TEST_CASE("Unit_hipCreateTextureObject_MmArrayResource") { + checkImageSupport(); + hipError_t ret; hipResourceDesc resDesc; hipTextureDesc texDesc; diff --git a/projects/hip-tests/catch/unit/texture/hipCreateTextureObject_Linear.cc b/projects/hip-tests/catch/unit/texture/hipCreateTextureObject_Linear.cc index 29bd35a821..0e2caf6a80 100644 --- a/projects/hip-tests/catch/unit/texture/hipCreateTextureObject_Linear.cc +++ b/projects/hip-tests/catch/unit/texture/hipCreateTextureObject_Linear.cc @@ -26,6 +26,8 @@ THE SOFTWARE. * Validates Linear Resource texture object with negative/functional tests. */ TEST_CASE("Unit_hipCreateTextureObject_LinearResource") { + checkImageSupport(); + float *texBuf; hipError_t ret; constexpr int xsize = 32; diff --git a/projects/hip-tests/catch/unit/texture/hipCreateTextureObject_Pitch2D.cc b/projects/hip-tests/catch/unit/texture/hipCreateTextureObject_Pitch2D.cc index 89eb936c3c..7fc593708c 100644 --- a/projects/hip-tests/catch/unit/texture/hipCreateTextureObject_Pitch2D.cc +++ b/projects/hip-tests/catch/unit/texture/hipCreateTextureObject_Pitch2D.cc @@ -29,6 +29,8 @@ THE SOFTWARE. * Validates Pitch2D Resource texture object with negative and functional tests */ TEST_CASE("Unit_hipCreateTextureObject_Pitch2DResource") { + checkImageSupport(); + hipError_t ret; hipResourceDesc resDesc; hipTextureDesc texDesc; diff --git a/projects/hip-tests/catch/unit/texture/hipTextureObjFetchVector.cc b/projects/hip-tests/catch/unit/texture/hipTextureObjFetchVector.cc index 8fc6c6730a..d3faa5d5b0 100644 --- a/projects/hip-tests/catch/unit/texture/hipTextureObjFetchVector.cc +++ b/projects/hip-tests/catch/unit/texture/hipTextureObjFetchVector.cc @@ -23,10 +23,12 @@ THE SOFTWARE. template __global__ void tex1dKernelFetch(T *val, hipTextureObject_t obj, int N) { +#if !defined(__HIP_NO_IMAGE_SUPPORT) || !__HIP_NO_IMAGE_SUPPORT int k = blockIdx.x * blockDim.x + threadIdx.x; if (k < N) { val[k] = tex1Dfetch(obj, k); } +#endif } template @@ -203,6 +205,8 @@ bool runTest(const char *description) { } TEST_CASE("Unit_hipTextureFetch_vector") { + checkImageSupport(); + // test for char runTest("char1"); runTest("char2"); diff --git a/projects/hip-tests/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp b/projects/hip-tests/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp index 613db7ff01..5f2ded8518 100644 --- a/projects/hip-tests/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp +++ b/projects/hip-tests/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp @@ -25,7 +25,9 @@ THE SOFTWARE. texture tex; extern "C" __global__ void tex2dKernel(float* outputData, int width, int height) { +#if !defined(__HIP_NO_IMAGE_SUPPORT) || !__HIP_NO_IMAGE_SUPPORT int x = hipBlockIdx_x * hipBlockDim_x + hipThreadIdx_x; int y = hipBlockIdx_y * hipBlockDim_y + hipThreadIdx_y; outputData[y * width + x] = tex2D(tex, x, y); +#endif } diff --git a/projects/hip-tests/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp b/projects/hip-tests/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp index 1e27b14008..161b3ea93d 100644 --- a/projects/hip-tests/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp +++ b/projects/hip-tests/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp @@ -127,7 +127,20 @@ bool runTest(int argc, char** argv) { return testResult; } +inline bool isImageSupported() { + int imageSupport = 1; +#ifdef __HIP_PLATFORM_AMD__ + HIPCHECK(hipDeviceGetAttribute(&imageSupport, hipDeviceAttributeImageSupport, + 0)); +#endif + return imageSupport != 0; +} + int main(int argc, char** argv) { + if (!isImageSupported()) { + printf("Texture is not support on the device. Skipped.\n"); + return 0; + } hipInit(0); testResult = runTest(argc, argv); printf("%s ...\n", testResult ? "PASSED" : "FAILED");