gdb/testsuite/gdb.rocm: Check value returned by hipDeviceSynchronize

Functions of the hip runtime returning a hipError_t can be marked nodiscard depending on the configuration[1] (when compiled with C++17). This patch makes sure that we always check the value returned by hipDeviceSynchronize and friends, and print an error message when appropriate. This avoid a wall of warnings when running the testsuite if the compiler defaults to using C++17. It is always a good practice to check the return values anyway. While is revisiting testsuite hip programs, the following changes are made: - The gdb.rocm/nonstop-mode.cpp hip application is cleaned-up since is not following GDB's coding standard. - The multi-GPU.exp testcase is removed because it is empty and therefore serves no purpose. [1] https://github.com/ROCm-Developer-Tools/HIP/blob/docs/5.7.1/include/hip/hip_runtime_api.h#L203-L218 (cherry picked from commit fded0fb) (cherry picked from commit 4457a19) (cherry picked from commit 5a2f58c) (cherry picked from commit 14bb3b9) Change-Id: I8397fe41d52fa7bd9997d787c2f36cc06e1bd9a8
ROCm · Oct 30, 2023 · b15405c · b15405c
1 parent 9fda287
commit b15405c
Show file tree

Hide file tree

Showing 28 changed files with 316 additions and 267 deletions.
diff --git a/gdb/testsuite/gdb.perf/rocm-break-cond-false.cpp b/gdb/testsuite/gdb.perf/rocm-break-cond-false.cpp
@@ -18,6 +18,17 @@
 
 #include <hip/hip_runtime.h>
 
+#define CHECK(cmd)                                                           \
+  {                                                                          \
+    hipError_t error = cmd;                                                  \
+    if (error != hipSuccess)                                                 \
+      {                                                                      \
+	fprintf (stderr, "error: '%s'(%d) at %s:%d\n",                       \
+		 hipGetErrorString (error), error, __FILE__, __LINE__);      \
+	exit (EXIT_FAILURE);                                                 \
+      }                                                                      \
+  }
+
 __global__ void
 kernel ()
 {
@@ -31,7 +42,7 @@ main (int argc, char **argv)
 		      0 /*dynamicShared*/, 0 /*stream*/);
 
   /* Wait until kernel finishes.  */
-  hipDeviceSynchronize ();
+  CHECK (hipDeviceSynchronize ());
 
   return 0;
 }
diff --git a/gdb/testsuite/gdb.rocm/aspace-watchpoint.cpp b/gdb/testsuite/gdb.rocm/aspace-watchpoint.cpp
@@ -26,6 +26,17 @@
 
 #include <hip/hip_runtime.h>
 
+#define CHECK(cmd)                                                           \
+  {                                                                          \
+    hipError_t error = cmd;                                                  \
+    if (error != hipSuccess)                                                 \
+      {                                                                      \
+	fprintf (stderr, "error: '%s'(%d) at %s:%d\n",                       \
+		 hipGetErrorString (error), error, __FILE__, __LINE__);      \
+	exit (EXIT_FAILURE);                                                 \
+      }                                                                      \
+  }
+
 __device__ void
 change_memory (char *private_ptr, int *global_ptr)
 {
@@ -74,11 +85,12 @@ int
 main (int argc, char* argv[])
 {
   int *global_ptr;
-  hipMalloc (&global_ptr, 4);
+  CHECK (hipMalloc (&global_ptr, 4));
   int init_value_h = 0;
-  hipMemcpy (global_ptr, &init_value_h, sizeof (init_value_h), hipMemcpyHostToDevice);
+  CHECK (hipMemcpy (global_ptr, &init_value_h, sizeof (init_value_h),
+		    hipMemcpyHostToDevice));
 
   kernel<<<1, 1, 0>>> (global_ptr);
-  hipDeviceSynchronize ();
+  CHECK (hipDeviceSynchronize ());
   return 0;
 }
diff --git a/gdb/testsuite/gdb.rocm/branch-fault.cpp b/gdb/testsuite/gdb.rocm/branch-fault.cpp
@@ -17,6 +17,17 @@
 
 #include <hip/hip_runtime.h>
 
+#define CHECK(cmd)                                                           \
+  {                                                                          \
+    hipError_t error = cmd;                                                  \
+    if (error != hipSuccess)                                                 \
+      {                                                                      \
+	fprintf (stderr, "error: '%s'(%d) at %s:%d\n",                       \
+		 hipGetErrorString (error), error, __FILE__, __LINE__);      \
+	exit (EXIT_FAILURE);                                                 \
+      }                                                                      \
+  }
+
 typedef void (*func_ptr) ();
 
 __global__ void
@@ -30,6 +41,6 @@ int
 main (int argc, char* argv[])
 {
   hipLaunchKernelGGL (kernel, dim3 (1), dim3 (1), 0, 0);
-  hipDeviceSynchronize ();
+  CHECK (hipDeviceSynchronize ());
   return 0;
 }
diff --git a/gdb/testsuite/gdb.rocm/breakpoint-after-exit.cpp b/gdb/testsuite/gdb.rocm/breakpoint-after-exit.cpp
@@ -19,6 +19,17 @@
 #include <unistd.h>
 #include <hip/hip_runtime.h>
 
+#define CHECK(cmd)                                                           \
+  {                                                                          \
+    hipError_t error = cmd;                                                  \
+    if (error != hipSuccess)                                                 \
+      {                                                                      \
+	fprintf (stderr, "error: '%s'(%d) at %s:%d\n",                       \
+		 hipGetErrorString (error), error, __FILE__, __LINE__);      \
+	exit (EXIT_FAILURE);                                                 \
+      }                                                                      \
+  }
+
 __global__ void
 the_kernel ()
 {}
@@ -27,7 +38,7 @@ int
 main ()
 {
   hipLaunchKernelGGL (the_kernel, dim3 (1), dim3 (1), 0, 0);
-  hipDeviceSynchronize (); /* set breakpoint here */
+  CHECK (hipDeviceSynchronize ()); /* set breakpoint here */
 
   return 0;
 }
diff --git a/gdb/testsuite/gdb.rocm/deep-stack.cpp b/gdb/testsuite/gdb.rocm/deep-stack.cpp
@@ -19,6 +19,17 @@
 #include <cstdio>
 #include <hip/hip_runtime.h>
 
+#define CHECK(cmd)                                                           \
+  {                                                                          \
+    hipError_t error = cmd;                                                  \
+    if (error != hipSuccess)                                                 \
+      {                                                                      \
+	fprintf (stderr, "error: '%s'(%d) at %s:%d\n",                       \
+		 hipGetErrorString (error), error, __FILE__, __LINE__);      \
+	exit (EXIT_FAILURE);                                                 \
+      }                                                                      \
+  }
+
 __device__ void
 base_case ()
 {
@@ -52,7 +63,7 @@ int
 main ()
 {
   hipLaunchKernelGGL (HIP_KERNEL_NAME (hip_deep), dim3 (1), dim3 (1), 0, 0);
-  hipDeviceSynchronize ();
+  CHECK (hipDeviceSynchronize ());
   return EXIT_SUCCESS;
 }
 
diff --git a/gdb/testsuite/gdb.rocm/device-interrupt.cpp b/gdb/testsuite/gdb.rocm/device-interrupt.cpp
@@ -23,6 +23,17 @@
 #include <unistd.h>
 #include <hip/hip_runtime.h>
 
+#define CHECK(cmd)                                                           \
+  {                                                                          \
+    hipError_t error = cmd;                                                  \
+    if (error != hipSuccess)                                                 \
+      {                                                                      \
+	fprintf (stderr, "error: '%s'(%d) at %s:%d\n",                       \
+		 hipGetErrorString (error), error, __FILE__, __LINE__);      \
+	exit (EXIT_FAILURE);                                                 \
+      }                                                                      \
+  }
+
 /* Kernel entry point.  Loop forever, while avoiding to peg the
    GPU.  */
 
@@ -52,7 +63,7 @@ main (int argc, char **argv)
 
   /* Wait until kernel finishes.  In this case, this blocks
      "forever".  */
-  hipDeviceSynchronize ();
+  CHECK (hipDeviceSynchronize ());
 
   return 0;
 }
diff --git a/gdb/testsuite/gdb.rocm/disassemble.cpp b/gdb/testsuite/gdb.rocm/disassemble.cpp
@@ -19,6 +19,17 @@
 
 #include <hip/hip_runtime.h>
 
+#define CHECK(cmd)                                                           \
+  {                                                                          \
+    hipError_t error = cmd;                                                  \
+    if (error != hipSuccess)                                                 \
+      {                                                                      \
+	fprintf (stderr, "error: '%s'(%d) at %s:%d\n",                       \
+		 hipGetErrorString (error), error, __FILE__, __LINE__);      \
+	exit (EXIT_FAILURE);                                                 \
+      }                                                                      \
+  }
+
 /* Kernel entry point. */
 __global__ void kernel ()
 {
@@ -33,7 +44,7 @@ main (int argc, char* argv[])
   hipLaunchKernelGGL (kernel, dim3 (1), dim3 (1),
       0 /*dynamicShared*/, 0 /*stream*/);
 
-  hipDeviceSynchronize ();
+  CHECK (hipDeviceSynchronize ());
 
   return 0;
 }
diff --git a/gdb/testsuite/gdb.rocm/displaced-stepping.cpp b/gdb/testsuite/gdb.rocm/displaced-stepping.cpp
@@ -19,6 +19,17 @@
 
 #include <hip/hip_runtime.h>
 
+#define CHECK(cmd)                                                           \
+  {                                                                          \
+    hipError_t error = cmd;                                                  \
+    if (error != hipSuccess)                                                 \
+      {                                                                      \
+	fprintf (stderr, "error: '%s'(%d) at %s:%d\n",                       \
+		 hipGetErrorString (error), error, __FILE__, __LINE__);      \
+	exit (EXIT_FAILURE);                                                 \
+      }                                                                      \
+  }
+
 __global__ void
 kernel ()
 {
@@ -49,7 +60,7 @@ main (int argc, char **argv)
 		      0 /*stream*/);
 
   /* Wait until kernel finishes.  */
-  hipDeviceSynchronize ();
+  CHECK (hipDeviceSynchronize ());
 
   return 0;
 }
diff --git a/gdb/testsuite/gdb.rocm/fork-exec-gpu-to-non-gpu-execer.cpp b/gdb/testsuite/gdb.rocm/fork-exec-gpu-to-non-gpu-execer.cpp
@@ -18,6 +18,17 @@
 #include <hip/hip_runtime.h>
 #include <unistd.h>
 
+#define CHECK(cmd)                                                           \
+  {                                                                          \
+    hipError_t error = cmd;                                                  \
+    if (error != hipSuccess)                                                 \
+      {                                                                      \
+	fprintf (stderr, "error: '%s'(%d) at %s:%d\n",                       \
+		 hipGetErrorString (error), error, __FILE__, __LINE__);      \
+	exit (EXIT_FAILURE);                                                 \
+      }                                                                      \
+  }
+
 __global__ static void
 kernel1 ()
 {}
@@ -50,6 +61,6 @@ main (int argc, char* argv[])
 
   hipLaunchKernelGGL (kernel2, dim3 (1), dim3 (1), 0, 0);
 
-  hipDeviceSynchronize ();
+  CHECK (hipDeviceSynchronize ());
   return 0;
 }
diff --git a/gdb/testsuite/gdb.rocm/fork-exec-non-gpu-to-gpu-execee.cpp b/gdb/testsuite/gdb.rocm/fork-exec-non-gpu-to-gpu-execee.cpp
@@ -17,6 +17,17 @@
 
 #include <hip/hip_runtime.h>
 
+#define CHECK(cmd)                                                           \
+  {                                                                          \
+    hipError_t error = cmd;                                                  \
+    if (error != hipSuccess)                                                 \
+      {                                                                      \
+	fprintf (stderr, "error: '%s'(%d) at %s:%d\n",                       \
+		 hipGetErrorString (error), error, __FILE__, __LINE__);      \
+	exit (EXIT_FAILURE);                                                 \
+      }                                                                      \
+  }
+
 __device__ static void
 break_here_execee ()
 {}
@@ -30,7 +41,7 @@ kernel ()
 int
 main (int argc, char* argv[])
 {
-  hipLaunchKernelGGL (kernel, dim3 (1), dim3 (1), 0, 0);
-  hipDeviceSynchronize ();
+  kernel<<<1, 1>>> ();
+  CHECK (hipDeviceSynchronize ());
   return 0;
 }
diff --git a/gdb/testsuite/gdb.rocm/lane-execution.cpp b/gdb/testsuite/gdb.rocm/lane-execution.cpp
@@ -137,7 +137,7 @@ main ()
 
   hipLaunchKernelGGL (kernel, grid_dim, block_dim, 0, 0, sInBuff, sOutBuff);
 
-  hipDeviceSynchronize ();
+  CHECK (hipDeviceSynchronize ());
 
   return 0;
 }
diff --git a/gdb/testsuite/gdb.rocm/lane-info.cpp b/gdb/testsuite/gdb.rocm/lane-info.cpp
@@ -19,6 +19,17 @@
 #include <hip/hip_runtime.h>
 #include <unistd.h>
 
+#define CHECK(cmd)                                                           \
+  {                                                                          \
+    hipError_t error = cmd;                                                  \
+    if (error != hipSuccess)                                                 \
+      {                                                                      \
+	fprintf (stderr, "error: '%s'(%d) at %s:%d\n",                       \
+		 hipGetErrorString (error), error, __FILE__, __LINE__);      \
+	exit (EXIT_FAILURE);                                                 \
+      }                                                                      \
+  }
+
 /* The kernel never returns, via this sleep, so that the .exp file can
    test background execution (cont&).  */
 
@@ -63,7 +74,7 @@ main ()
   hipLaunchKernelGGL (kernel, dim3 (1), dim3 (64 + 5),
 		      0 /*dynamicShared*/, 0 /*stream*/);
 
-  hipDeviceSynchronize ();
+  CHECK (hipDeviceSynchronize ());
 
   return 0;
 }
diff --git a/gdb/testsuite/gdb.rocm/line-breakpoint-in-kernel.cpp b/gdb/testsuite/gdb.rocm/line-breakpoint-in-kernel.cpp
@@ -18,6 +18,17 @@
 
 #include <hip/hip_runtime.h>
 
+#define CHECK(cmd)                                                           \
+  {                                                                          \
+    hipError_t error = cmd;                                                  \
+    if (error != hipSuccess)                                                 \
+      {                                                                      \
+	fprintf (stderr, "error: '%s'(%d) at %s:%d\n",                       \
+		 hipGetErrorString (error), error, __FILE__, __LINE__);      \
+	exit (EXIT_FAILURE);                                                 \
+      }                                                                      \
+  }
+
 __global__ void
 kernel ()
 {
@@ -31,7 +42,7 @@ main (int argc, char **argv)
   hipLaunchKernelGGL (kernel, dim3 (1), dim3 (1), 0, 0);
 
   /* Wait until kernel finishes.  */
-  hipDeviceSynchronize ();
+  CHECK (hipDeviceSynchronize ());
 
   return 0;
 }
diff --git a/gdb/testsuite/gdb.rocm/mi-aspace.cpp b/gdb/testsuite/gdb.rocm/mi-aspace.cpp
@@ -18,6 +18,17 @@
 
 #include <hip/hip_runtime.h>
 
+#define CHECK(cmd)                                                           \
+  {                                                                          \
+    hipError_t error = cmd;                                                  \
+    if (error != hipSuccess)                                                 \
+      {                                                                      \
+	fprintf (stderr, "error: '%s'(%d) at %s:%d\n",                       \
+		 hipGetErrorString (error), error, __FILE__, __LINE__);      \
+	exit (EXIT_FAILURE);                                                 \
+      }                                                                      \
+  }
+
 int global_var;
 
 __device__ void
@@ -41,7 +52,7 @@ main ()
      local#0.  */
   hipLaunchKernelGGL (kernel, grid_dim, block_dim, 1024, 0);
 
-  hipDeviceSynchronize ();
+  CHECK (hipDeviceSynchronize ());
 
   return 0;
 }