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: 70ccfa39ad]

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; }

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;

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;

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;

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;

projects/hip-tests/catch/unit/texture/hipTextureObjFetchVector.cc

@@ -23,10 +23,12 @@ THE SOFTWARE.
 
 template <typename T>
 __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<T>(obj, k);
   }
+#endif
 }
 
 template <typename T>
@@ -203,6 +205,8 @@ bool runTest(const char *description) {
 }
 
 TEST_CASE("Unit_hipTextureFetch_vector") {
+  checkImageSupport();
+
   // test for char
   runTest<char1>("char1");
   runTest<char2>("char2");

projects/hip-tests/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp

@@ -25,7 +25,9 @@ THE SOFTWARE.
 texture<float, 2, hipReadModeElementType> 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
 }

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");