Skip to content

Commit 12627cf

Browse files
music-dinomusic-dino
committed
Merge branch 'utils' into hipMemAdvise_tests
2 parents 8ddd164 + 09ce86a commit 12627cf

File tree

99 files changed

+207
-44
lines changed

Some content is hidden

Large Commits have some content hidden by default. Use the searchbox below for content that may be hidden.

99 files changed

+207
-44
lines changed

‎tests/catch/ABM/AddKernels/add.cc‎

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -29,6 +29,7 @@ TEMPLATE_TEST_CASE("ABM_AddKernel_MultiTypeMultiSize", "", int, long, float, lon
2929
REQUIRE(res == hipSuccess);
3030

3131
hipLaunchKernelGGL(add<TestType>, 1, size, 0, 0, d_a, d_b, d_c, size);
32+
HIP_CHECK(hipGetLastError());
3233

3334
res = hipMemcpy(a.data(), d_c, sizeof(TestType) * size, hipMemcpyDeviceToHost);
3435
REQUIRE(res == hipSuccess);

‎tests/catch/TypeQualifiers/hipManagedKeyword.cc‎

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -53,6 +53,7 @@ TEST_CASE("Unit_hipManagedKeyword_SingleGpu") {
5353
hipLaunchKernelGGL(add, dimGrid, dimBlock, 0, 0, static_cast<const float*>(A),
5454
static_cast<float*>(B));
5555

56+
HIP_CHECK(hipGetLastError());
5657
HIP_CHECK(hipDeviceSynchronize());
5758

5859
float maxError = 0.0f;

‎tests/catch/include/hip_test_checkers.hh‎

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -26,7 +26,7 @@ THE SOFTWARE.
2626
#include <fstream>
2727
#include <regex>
2828
#include <type_traits>
29-
29+
#define TOL 0.001
3030
#define guarantee(cond, str) \
3131
{ \
3232
if (!(cond)) { \
@@ -45,7 +45,7 @@ size_t checkVectors(T* A, T* B, T* Out, size_t N, T (*F)(T a, T b), bool expectM
4545
size_t mismatchesToPrint = 10;
4646
for (size_t i = 0; i < N; i++) {
4747
T expected = F(A[i], B[i]);
48-
if (Out[i] != expected) {
48+
if (std::fabs(Out[i] - expected) > TOL) {
4949
if (mismatchCount == 0) {
5050
firstMismatch = i;
5151
}

‎tests/catch/include/hip_test_common.hh‎

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -270,6 +270,7 @@ void launchKernel(K kernel, Dim numBlocks, Dim numThreads, std::uint32_t memPerB
270270
launchRTCKernel<Typenames...>(kernel, numBlocks, numThreads, memPerBlock, stream,
271271
std::forward<Args>(packedArgs)...);
272272
#endif
273+
HIP_CHECK(hipGetLastError());
273274
}
274275

275276
//---

‎tests/catch/include/resource_guards.hh‎

Lines changed: 5 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -67,15 +67,16 @@ template <typename T> class LinearAllocGuard {
6767
free(ptr_);
6868
break;
6969
case LinearAllocs::mallocAndRegister:
70-
hipHostUnregister(host_ptr_);
70+
// Cast to void to suppress nodiscard warnings
71+
static_cast<void>(hipHostUnregister(host_ptr_));
7172
free(host_ptr_);
7273
break;
7374
case LinearAllocs::hipHostMalloc:
74-
hipHostFree(ptr_);
75+
static_cast<void>(hipHostFree(ptr_));
7576
break;
7677
case LinearAllocs::hipMalloc:
7778
case LinearAllocs::hipMallocManaged:
78-
hipFree(ptr_);
79+
static_cast<void>(hipFree(ptr_));
7980
}
8081
}
8182

@@ -112,7 +113,7 @@ class StreamGuard {
112113

113114
~StreamGuard() {
114115
if (stream_type_ == Streams::created) {
115-
hipStreamDestroy(stream_);
116+
static_cast<void>(hipStreamDestroy(stream_));
116117
}
117118
}
118119

‎tests/catch/include/utils.hh‎

Lines changed: 19 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -38,6 +38,7 @@ void ArrayMismatch(T* const expected, T* const actual, const size_t num_elements
3838
}
3939
}
4040

41+
<<<<<<< HEAD
4142
template <typename T>
4243
void ArrayFindIfNot(T* const array, const T expected_value, const size_t num_elements) {
4344
const auto it = std::find_if_not(array, array + num_elements, [expected_value](const int elem) {
@@ -48,10 +49,28 @@ void ArrayFindIfNot(T* const array, const T expected_value, const size_t num_ele
4849
const auto idx = std::distance(array, it);
4950
INFO("Value mismatch at index " << idx);
5051
REQUIRE(expected_value == array[idx]);
52+
=======
53+
template <typename It, typename T> void ArrayFindIfNot(It begin, It end, const T expected_value) {
54+
const auto it = std::find_if_not(
55+
begin, end, [expected_value](const int elem) { return expected_value == elem; });
56+
57+
if (it != end) {
58+
const auto idx = std::distance(begin, it);
59+
INFO("Value mismatch at index " << idx);
60+
REQUIRE(expected_value == *it);
61+
>>>>>>> utils
5162
}
5263
}
5364

5465
template <typename T>
66+
<<<<<<< HEAD
67+
=======
68+
void ArrayFindIfNot(T* const array, const T expected_value, const size_t num_elements) {
69+
ArrayFindIfNot(array, array + num_elements, expected_value);
70+
}
71+
72+
template <typename T>
73+
>>>>>>> utils
5574
__global__ void VectorIncrement(T* const vec, const T increment_value, size_t N) {
5675
size_t offset = (blockIdx.x * blockDim.x + threadIdx.x);
5776
size_t stride = blockDim.x * gridDim.x;

‎tests/catch/multiproc/hipIpcEventHandle.cc‎

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -221,6 +221,7 @@ void runMultiProcKernel(ipcEventInfo_t *shmEventInfo, int index) {
221221
const dim3 blocks(BUF_SIZE / threads.x, 1);
222222
hipLaunchKernelGGL(computeKernel, dim3(blocks), dim3(threads), 0, 0,
223223
d_ptr + index *BUF_SIZE, d_ptr, index + 1);
224+
HIP_CHECK(hipGetLastError());
224225
HIP_CHECK(hipEventRecord(event));
225226

226227
// Barrier 2 : Signals that event is recorded

‎tests/catch/multiproc/hipMallocConcurrencyMproc.cc‎

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -120,7 +120,7 @@ static bool validateMemoryOnGPU(int gpu, bool concurOnOneGPU = false) {
120120
hipLaunchKernelGGL(HipTest::vectorADD, dim3(blocks), dim3(threadsPerBlock),
121121
0, 0, static_cast<const int*>(A_d),
122122
static_cast<const int*>(B_d), C_d, N);
123-
123+
HIP_CHECK(hipGetLastError());
124124
HIP_CHECK(hipMemcpy(C_h, C_d, Nbytes, hipMemcpyDeviceToHost));
125125

126126
if (!HipTest::checkVectorADD(A_h, B_h, C_h, N)) {

‎tests/catch/stress/printf/Stress_printf_ComplexKernels.cc‎

Lines changed: 6 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -276,12 +276,15 @@ bool test_printf_multistream(uint32_t num_blocks,
276276
hipLaunchKernelGGL(kernel_complex_opX, dim3(num_blocks, 1, 1),
277277
dim3(threads_per_block, 1, 1),
278278
0, stream[i], Ad, Bd, iterCount);
279+
HIP_CHECK(hipGetLastError());
279280
hipLaunchKernelGGL(kernel_complex_opY, dim3(1, num_blocks, 1),
280281
dim3(1, threads_per_block, 1),
281282
0, stream[i], Ad, Bd, iterCount);
283+
HIP_CHECK(hipGetLastError());
282284
hipLaunchKernelGGL(kernel_complex_opZ, dim3(1, 1, num_blocks),
283285
dim3(1, 1, threads_per_block),
284286
0, stream[i], Ad, Bd, iterCount);
287+
HIP_CHECK(hipGetLastError());
285288
}
286289
HIP_CHECK(hipDeviceSynchronize());
287290
for (int i = 0; i < NUM_STREAM; i++) {
@@ -368,12 +371,15 @@ bool test_printf_multigpu(int gpu,
368371
hipLaunchKernelGGL(kernel_complex_opX, dim3(num_blocks, 1, 1),
369372
dim3(threads_per_block, 1, 1),
370373
0, 0, Ad, Bd, iterCount);
374+
HIP_CHECK(hipGetLastError());
371375
hipLaunchKernelGGL(kernel_complex_opY, dim3(1, num_blocks, 1),
372376
dim3(1, threads_per_block, 1),
373377
0, 0, Ad, Bd, iterCount);
378+
HIP_CHECK(hipGetLastError());
374379
hipLaunchKernelGGL(kernel_complex_opZ, dim3(1, 1, num_blocks),
375380
dim3(1, 1, threads_per_block),
376381
0, 0, Ad, Bd, iterCount);
382+
HIP_CHECK(hipGetLastError());
377383
HIP_CHECK(hipDeviceSynchronize());
378384
std::ifstream CapturedData = captured.getCapturedData();
379385
char *buffer = new char[CHUNK_SIZE];

‎tests/catch/stress/printf/Stress_printf_SimpleKernels.cc‎

Lines changed: 5 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -246,6 +246,7 @@ bool test_printf_conststr(uint32_t num_blocks, uint32_t threads_per_block,
246246
hipLaunchKernelGGL(kernel_printf_conststr, dim3(num_blocks, 1, 1),
247247
dim3(threads_per_block, 1, 1),
248248
0, 0, iterCount);
249+
HIP_CHECK(hipGetLastError());
249250
HIP_CHECK(hipStreamSynchronize(0));
250251
std::ifstream CapturedData = captured.getCapturedData();
251252
char *buffer = new char[CHUNK_SIZE];
@@ -308,6 +309,7 @@ bool test_printf_two_conditionalstr(uint32_t num_blocks,
308309
dim3(num_blocks, 1, 1),
309310
dim3(threads_per_block, 1, 1),
310311
0, 0, iterCount);
312+
HIP_CHECK(hipGetLastError());
311313
HIP_CHECK(hipStreamSynchronize(0));
312314
std::ifstream CapturedData = captured.getCapturedData();
313315
char *buffer = new char[CHUNK_SIZE];
@@ -370,6 +372,7 @@ bool test_printf_single_conditionalstr(uint32_t num_blocks,
370372
dim3(num_blocks, 1, 1),
371373
dim3(threads_per_block, 1, 1),
372374
0, 0, iterCount);
375+
HIP_CHECK(hipGetLastError());
373376
HIP_CHECK(hipStreamSynchronize(0));
374377
std::ifstream CapturedData = captured.getCapturedData();
375378
char *buffer = new char[CHUNK_SIZE];
@@ -427,6 +430,7 @@ bool test_variable_str(uint32_t print_limit,
427430
hipLaunchKernelGGL(func, dim3(num_blocks, 1, 1),
428431
dim3(threads_per_block, 1, 1),
429432
0, 0, iterCount, Ad);
433+
HIP_CHECK(hipGetLastError());
430434
HIP_CHECK(hipStreamSynchronize(0));
431435
HIP_CHECK(hipMemcpy(Ah, Ad, buffsize*sizeof(int32_t),
432436
hipMemcpyDeviceToHost));
@@ -483,6 +487,7 @@ bool test_decimal_str(uint32_t num_blocks, uint32_t threads_per_block,
483487
hipLaunchKernelGGL(kernel_decimal_calculation, dim3(num_blocks, 1, 1),
484488
dim3(threads_per_block, 1, 1),
485489
0, 0, iterCount, maxPrecision);
490+
HIP_CHECK(hipGetLastError());
486491
HIP_CHECK(hipStreamSynchronize(0));
487492
std::ifstream CapturedData = captured.getCapturedData();
488493
char *buffer = new char[CHUNK_SIZE];

0 commit comments

Comments
 (0)