AI/rocm-systems
2
0
Garpu 0
Kode Masalah Tarik Permintaan Actions Packages Projects Rilis Wiki Kegiatan

SWDEV-425248 - numerous bug fixes

Telusuri Sumber 
https://gerrit-git.amd.com/c/compute/ec/clr/+/849086 changed stream sync behaviour and accordingly updated Unit_hipStreamCreateWithFlags_DefaultStreamInteraction.
Update error code to match for hipStreamWaitEvent.
Thread detach makes executions all async and while that is happening, It is unexpected effect for device reset so join thread is better.

Change-Id: I1affa84089626dee478d8bcc5aaa318e320fd6b0


[ROCm/hip-tests commit: be88e1dffc]
This commit is contained in:
Jaydeep Patel
2023-10-17 17:25:59 +00:00
committed by Rakesh Roy
orang tua e96f828db3
melakukan d4e5dfa942
3 mengubah file dengan 10 tambahan dan 7 penghapusan
Download Patch File Download Diff File Expand all files Collapse all files
projects/hip-tests/catch/unit/stream/hipStreamCreateWithFlags.cc
+5 -5
Melihat File
@@ -59,10 +59,7 @@ TEST_CASE("Unit_hipStreamCreateWithFlags_Default") {
TEST_CASE("Unit_hipStreamCreateWithFlags_DefaultStreamInteraction") { TEST_CASE("Unit_hipStreamCreateWithFlags_DefaultStreamInteraction") {
const hipStream_t defaultStream = GENERATE(static_cast<hipStream_t>(nullptr), hipStreamPerThread); const hipStream_t defaultStream = GENERATE(static_cast<hipStream_t>(nullptr), hipStreamPerThread);
const unsigned int flagUnderTest = GENERATE(hipStreamDefault, hipStreamNonBlocking); const unsigned int flagUnderTest = GENERATE(hipStreamDefault, hipStreamNonBlocking);
const hipError_t expectedError = (flagUnderTest == hipStreamDefault) && (defaultStream == nullptr) CAPTURE(defaultStream, flagUnderTest);
? hipErrorNotReady
: hipSuccess;
CAPTURE(defaultStream, flagUnderTest, expectedError, hipGetErrorString(expectedError));
hipStream_t stream{}; hipStream_t stream{};
HIP_CHECK(hipStreamCreateWithFlags(&stream, flagUnderTest)); HIP_CHECK(hipStreamCreateWithFlags(&stream, flagUnderTest));
@@ -70,12 +67,15 @@ TEST_CASE("Unit_hipStreamCreateWithFlags_DefaultStreamInteraction") {
constexpr auto delay = std::chrono::milliseconds(500); constexpr auto delay = std::chrono::milliseconds(500);
SECTION("default stream waiting for created stream") { SECTION("default stream waiting for created stream") {
const hipError_t expectedError = (flagUnderTest == hipStreamDefault) && (defaultStream == nullptr)
? hipErrorNotReady
: hipSuccess;
LaunchDelayKernel(delay, stream); LaunchDelayKernel(delay, stream);
REQUIRE(hipStreamQuery(defaultStream) == expectedError); REQUIRE(hipStreamQuery(defaultStream) == expectedError);
} }
SECTION("created stream waiting for default stream") { SECTION("created stream waiting for default stream") {
LaunchDelayKernel(delay, defaultStream); LaunchDelayKernel(delay, defaultStream);
REQUIRE(hipStreamQuery(stream) == expectedError); REQUIRE(hipStreamQuery(stream) == hipSuccess);
} }
HIP_CHECK(hipDeviceSynchronize()); HIP_CHECK(hipDeviceSynchronize());
projects/hip-tests/catch/unit/stream/hipStreamWaitEvent.cc
+1 -1
Melihat File
@@ -73,7 +73,7 @@ TEST_CASE("Unit_hipStreamWaitEvent_UninitializedStream_Negative") {
HIP_CHECK(hipEventCreate(&event)); HIP_CHECK(hipEventCreate(&event));
HIP_CHECK_ERROR(hipStreamWaitEvent(stream, event, 0), hipErrorContextIsDestroyed); HIP_CHECK_ERROR(hipStreamWaitEvent(stream, event, 0), hipErrorInvalidHandle);
HIP_CHECK(hipEventDestroy(event)); HIP_CHECK(hipEventDestroy(event));
} }
projects/hip-tests/catch/unit/streamperthread/hipStreamPerThread_DeviceReset.cc
+4 -1
Melihat File
@@ -51,8 +51,11 @@ TEST_CASE("Unit_hipStreamPerThread_DeviceReset_1") {
for (auto &th : threads) { for (auto &th : threads) {
th = std::thread(Copy_to_device); th = std::thread(Copy_to_device);
th.detach();
} }
for (auto &th : threads) {
th.join();
}
HIP_CHECK(hipDeviceReset()); HIP_CHECK(hipDeviceReset());
} }