pivot test fix (#65)

* adding device sync in spin loop tests to not overwrite pivots before checking them * bump version
ROCm · May 6, 2020 · e1c0144 · e1c0144
1 parent a1a5ae0
commit e1c0144
Show file tree

Hide file tree

Showing 5 changed files with 49 additions and 1 deletion.
diff --git a/CMakeLists.txt b/CMakeLists.txt
@@ -85,7 +85,7 @@ option(BUILD_VERBOSE "Output additional build information" OFF)
 include(cmake/Dependencies.cmake)
 
 # Setup version
-rocm_setup_version(VERSION 1.12.9)
+rocm_setup_version(VERSION 1.12.10)
 set(rocsparse_SOVERSION 0.1)
 
 # AMD targets

diff --git a/clients/include/testing_csric0.hpp b/clients/include/testing_csric0.hpp
@@ -496,6 +496,9 @@ void testing_csric0(const Arguments& arg)
                                 (h_analysis_pivot_1[0] != -1) ? rocsparse_status_zero_pivot
                                                               : rocsparse_status_success);
 
+        // Sync to force updated pivots
+        CHECK_HIP_ERROR(hipDeviceSynchronize());
+
         // Pointer mode device
         CHECK_ROCSPARSE_ERROR(rocsparse_set_pointer_mode(handle, rocsparse_pointer_mode_device));
         CHECK_ROCSPARSE_ERROR(rocsparse_csric0_analysis<T>(handle,
@@ -513,6 +516,9 @@ void testing_csric0(const Arguments& arg)
                                 (h_analysis_pivot_1[0] != -1) ? rocsparse_status_zero_pivot
                                                               : rocsparse_status_success);
 
+        // Sync to force updated pivots
+        CHECK_HIP_ERROR(hipDeviceSynchronize());
+
         // Perform solve step
 
         // Pointer mode host
@@ -523,6 +529,9 @@ void testing_csric0(const Arguments& arg)
                                 (h_solve_pivot_1[0] != -1) ? rocsparse_status_zero_pivot
                                                            : rocsparse_status_success);
 
+        // Sync to force updated pivots
+        CHECK_HIP_ERROR(hipDeviceSynchronize());
+
         // Pointer mode device
         CHECK_ROCSPARSE_ERROR(rocsparse_set_pointer_mode(handle, rocsparse_pointer_mode_device));
         CHECK_ROCSPARSE_ERROR(rocsparse_csric0<T>(
@@ -531,6 +540,9 @@ void testing_csric0(const Arguments& arg)
                                 (h_solve_pivot_1[0] != -1) ? rocsparse_status_zero_pivot
                                                            : rocsparse_status_success);
 
+        // Sync to force updated pivots
+        CHECK_HIP_ERROR(hipDeviceSynchronize());
+
         // Copy output to host
         CHECK_HIP_ERROR(hipMemcpy(hcsr_val_1, dcsr_val_1, sizeof(T) * nnz, hipMemcpyDeviceToHost));
         CHECK_HIP_ERROR(hipMemcpy(hcsr_val_2, dcsr_val_2, sizeof(T) * nnz, hipMemcpyDeviceToHost));

diff --git a/clients/include/testing_csrilu0.hpp b/clients/include/testing_csrilu0.hpp
@@ -501,6 +501,9 @@ void testing_csrilu0(const Arguments& arg)
                                 (h_analysis_pivot_1[0] != -1) ? rocsparse_status_zero_pivot
                                                               : rocsparse_status_success);
 
+        // Sync to force updated pivots
+        CHECK_HIP_ERROR(hipDeviceSynchronize());
+
         // Pointer mode device
         CHECK_ROCSPARSE_ERROR(rocsparse_set_pointer_mode(handle, rocsparse_pointer_mode_device));
         CHECK_ROCSPARSE_ERROR(rocsparse_csrilu0_analysis<T>(handle,
@@ -518,6 +521,9 @@ void testing_csrilu0(const Arguments& arg)
                                 (h_analysis_pivot_1[0] != -1) ? rocsparse_status_zero_pivot
                                                               : rocsparse_status_success);
 
+        // Sync to force updated pivots
+        CHECK_HIP_ERROR(hipDeviceSynchronize());
+
         // Perform solve step
 
         // Pointer mode host
@@ -528,6 +534,9 @@ void testing_csrilu0(const Arguments& arg)
                                 (h_solve_pivot_1[0] != -1) ? rocsparse_status_zero_pivot
                                                            : rocsparse_status_success);
 
+        // Sync to force updated pivots
+        CHECK_HIP_ERROR(hipDeviceSynchronize());
+
         // Pointer mode device
         CHECK_ROCSPARSE_ERROR(rocsparse_set_pointer_mode(handle, rocsparse_pointer_mode_device));
         CHECK_ROCSPARSE_ERROR(rocsparse_csrilu0<T>(
@@ -536,6 +545,9 @@ void testing_csrilu0(const Arguments& arg)
                                 (h_solve_pivot_1[0] != -1) ? rocsparse_status_zero_pivot
                                                            : rocsparse_status_success);
 
+        // Sync to force updated pivots
+        CHECK_HIP_ERROR(hipDeviceSynchronize());
+
         // Copy output to host
         CHECK_HIP_ERROR(hipMemcpy(hcsr_val_1, dcsr_val_1, sizeof(T) * nnz, hipMemcpyDeviceToHost));
         CHECK_HIP_ERROR(hipMemcpy(hcsr_val_2, dcsr_val_2, sizeof(T) * nnz, hipMemcpyDeviceToHost));

diff --git a/clients/include/testing_csrsm.hpp b/clients/include/testing_csrsm.hpp
@@ -809,6 +809,9 @@ void testing_csrsm(const Arguments& arg)
                                 (h_analysis_pivot_1[0] != -1) ? rocsparse_status_zero_pivot
                                                               : rocsparse_status_success);
 
+        // Sync to force updated pivots
+        CHECK_HIP_ERROR(hipDeviceSynchronize());
+
         // Pointer mode device
         CHECK_ROCSPARSE_ERROR(rocsparse_set_pointer_mode(handle, rocsparse_pointer_mode_device));
         CHECK_ROCSPARSE_ERROR(rocsparse_csrsm_analysis<T>(handle,
@@ -832,6 +835,9 @@ void testing_csrsm(const Arguments& arg)
                                 (h_analysis_pivot_1[0] != -1) ? rocsparse_status_zero_pivot
                                                               : rocsparse_status_success);
 
+        // Sync to force updated pivots
+        CHECK_HIP_ERROR(hipDeviceSynchronize());
+
         // Perform solve step
 
         // Pointer mode host
@@ -856,6 +862,9 @@ void testing_csrsm(const Arguments& arg)
                                 (h_solve_pivot_1[0] != -1) ? rocsparse_status_zero_pivot
                                                            : rocsparse_status_success);
 
+        // Sync to force updated pivots
+        CHECK_HIP_ERROR(hipDeviceSynchronize());
+
         // Pointer mode device
         CHECK_ROCSPARSE_ERROR(rocsparse_set_pointer_mode(handle, rocsparse_pointer_mode_device));
         CHECK_ROCSPARSE_ERROR(rocsparse_csrsm_solve<T>(handle,
@@ -878,6 +887,9 @@ void testing_csrsm(const Arguments& arg)
                                 (h_solve_pivot_1[0] != -1) ? rocsparse_status_zero_pivot
                                                            : rocsparse_status_success);
 
+        // Sync to force updated pivots
+        CHECK_HIP_ERROR(hipDeviceSynchronize());
+
         // Copy output to host
         CHECK_HIP_ERROR(hipMemcpy(hB_1, dB_1, sizeof(T) * M * nrhs, hipMemcpyDeviceToHost));
         CHECK_HIP_ERROR(hipMemcpy(hB_2, dB_2, sizeof(T) * M * nrhs, hipMemcpyDeviceToHost));

diff --git a/clients/include/testing_csrsv.hpp b/clients/include/testing_csrsv.hpp
@@ -631,6 +631,9 @@ void testing_csrsv(const Arguments& arg)
                                 (h_analysis_pivot_1[0] != -1) ? rocsparse_status_zero_pivot
                                                               : rocsparse_status_success);
 
+        // Sync to force updated pivots
+        CHECK_HIP_ERROR(hipDeviceSynchronize());
+
         // Pointer mode device
         CHECK_ROCSPARSE_ERROR(rocsparse_set_pointer_mode(handle, rocsparse_pointer_mode_device));
         CHECK_ROCSPARSE_ERROR(rocsparse_csrsv_analysis<T>(handle,
@@ -649,6 +652,9 @@ void testing_csrsv(const Arguments& arg)
                                 (h_analysis_pivot_1[0] != -1) ? rocsparse_status_zero_pivot
                                                               : rocsparse_status_success);
 
+        // Sync to force updated pivots
+        CHECK_HIP_ERROR(hipDeviceSynchronize());
+
         // Perform solve step
 
         // Pointer mode host
@@ -671,6 +677,9 @@ void testing_csrsv(const Arguments& arg)
                                 (h_solve_pivot_1[0] != -1) ? rocsparse_status_zero_pivot
                                                            : rocsparse_status_success);
 
+        // Sync to force updated pivots
+        CHECK_HIP_ERROR(hipDeviceSynchronize());
+
         // Pointer mode device
         CHECK_ROCSPARSE_ERROR(rocsparse_set_pointer_mode(handle, rocsparse_pointer_mode_device));
         CHECK_ROCSPARSE_ERROR(rocsparse_csrsv_solve<T>(handle,
@@ -691,6 +700,9 @@ void testing_csrsv(const Arguments& arg)
                                 (h_solve_pivot_1[0] != -1) ? rocsparse_status_zero_pivot
                                                            : rocsparse_status_success);
 
+        // Sync to force updated pivots
+        CHECK_HIP_ERROR(hipDeviceSynchronize());
+
         // Copy output to host
         CHECK_HIP_ERROR(hipMemcpy(hy_1, dy_1, sizeof(T) * M, hipMemcpyDeviceToHost));
         CHECK_HIP_ERROR(hipMemcpy(hy_2, dy_2, sizeof(T) * M, hipMemcpyDeviceToHost));