Skip to content

Commit

Permalink
SWDEV-399632/SWDEV-403960 - [catch2][dtest] Enable xnack+ check condi…
Browse files Browse the repository at this point in the history 
…tion (#364)

Change-Id: I9627d75d0d3258cf261c8e4bfe6c7c3c35c8f9c1
  • Loading branch information
@rocm-ci @gargrahul
rocm-ci authored and gargrahul committed Aug 10, 2023
1 parent b1a5fae commit 9cf8b32
Show file tree
Hide file tree
Showing 3 changed files with 149 additions and 265 deletions.
307 changes: 43 additions & 264 deletions catch/multiproc/hipMemCoherencyTstMProc.cc
View file
Original file line number Diff line number Diff line change
Expand Up @@ -81,13 +81,16 @@ static void TstCoherency(int *Ptr, bool HmmMem) {

// Getting gpu frequency
if (IsGfx11()) {
HIPCHECK(hipDeviceGetAttribute(&peak_clk, hipDeviceAttributeWallClockRate, 0));
HIPCHECK(hipDeviceGetAttribute(&peak_clk,
hipDeviceAttributeWallClockRate, 0));
} else {
HIPCHECK(hipDeviceGetAttribute(&peak_clk, hipDeviceAttributeClockRate, 0));
HIPCHECK(hipDeviceGetAttribute(&peak_clk,
hipDeviceAttributeClockRate, 0));
}

if (!HmmMem) {
HIP_CHECK(hipHostGetDevicePointer(reinterpret_cast<void **>(&Dptr), Ptr, 0));
HIP_CHECK(hipHostGetDevicePointer(reinterpret_cast<void **>(&Dptr),
Ptr, 0));
if (IsGfx11()) {
CoherentTst_gfx11<<<1, 1, 0, strm>>>(Dptr, peak_clk);
} else {
Expand Down Expand Up @@ -122,58 +125,13 @@ static void TstCoherency(int *Ptr, bool HmmMem) {
// The following test is failing on Nvidia platform hence disabled it for now
#if HT_AMD
TEST_CASE("Unit_malloc_CoherentTst") {
if ((setenv("HSA_XNACK", "1", 1)) != 0) {
WARN("Unable to turn on HSA_XNACK, hence terminating the Test case!");
REQUIRE(false);
}
// The following code block is used to check for gfx906/8 so as to skip if
// any of the gpus available
int fd1[2]; // Used to store two ends of first pipe
pid_t p;
if (pipe(fd1) == -1) {
fprintf(stderr, "Pipe Failed");
REQUIRE(false);
}

/* GpuId[0] for gfx90a exists--> 1 for yes and 0 for no*/
int GpuId[1] = {0};
p = fork();

if (p < 0) {
fprintf(stderr, "fork Failed");
REQUIRE(false);
} else if (p > 0) { // parent process
close(fd1[1]); // Close writing end of first pipe
// Wait for child to send a string
wait(NULL);
// Read string from child and close reading end.
read(fd1[0], GpuId, 2 * sizeof(int));
close(fd1[0]);
if (GpuId[0] == 0) {
WARN("This test is applicable for MI200."
"Skipping the test!!");
exit(0);
}
} else { // child process
close(fd1[0]); // Close read end of first pipe
hipDeviceProp_t prop;
HIPCHECK(hipGetDeviceProperties(&prop, 0));
char *p = NULL;

if (CheckIfFeatSupported(CTFeatures::CT_FEATURE_FINEGRAIN_HWSUPPORT, prop.gcnArchName)) {
WARN("gfx90a gpu found on this system!!");
GpuId[0] = 1;
}

// Write concatenated string and close writing end
write(fd1[1], GpuId, 2 * sizeof(int));
close(fd1[1]);
exit(0);
}

// Test Case execution begins from here
int stat = 0;
if (fork() == 0) {
hipDeviceProp_t prop;
HIPCHECK(hipGetDeviceProperties(&prop, 0));
char *p = NULL;
p = strstr(prop.gcnArchName, "xnack+");
if (p) {
// Test Case execution begins from here
int stat = 0;
int managed = 0;
HIPCHECK(hipDeviceGetAttribute(&managed, hipDeviceAttributeManagedMemory,
0));
Expand All @@ -185,23 +143,10 @@ TEST_CASE("Unit_malloc_CoherentTst") {
Ptr = reinterpret_cast<int*>(malloc(SIZE));
TstCoherency(Ptr, HmmMem);
free(Ptr);
if (YES_COHERENT) {
// exit() with code 10 which indicates pass
exit(10);
} else {
// exit() with code 9 which indicates fail
exit(9);
}
} else {
SUCCEED("GPU 0 doesn't support hipDeviceAttributeManagedMemory "
"attribute. Hence skipping the testing with Pass result.\n");
}
REQUIRE(YES_COHERENT);
}
} else {
wait(&stat);
int Result = WEXITSTATUS(stat);
if (Result != 10) {
REQUIRE(false);
}
HipTest::HIP_SKIP_TEST("GPU is not xnack enabled hence skipping the test...\n");
}
}
#endif
Expand All @@ -212,55 +157,12 @@ TEST_CASE("Unit_malloc_CoherentTst") {
// The following test is failing on Nvidia platform hence disabling it for now
#if HT_AMD
TEST_CASE("Unit_malloc_CoherentTstWthAdvise") {
if ((setenv("HSA_XNACK", "1", 1)) != 0) {
WARN("Unable to turn on HSA_XNACK, hence terminating the Test case!");
REQUIRE(false);
}
// The following code block is used to check for gfx906/8 so as to skip if
// any of the gpus available
int fd1[2]; // Used to store two ends of first pipe
pid_t p;
if (pipe(fd1) == -1) {
fprintf(stderr, "Pipe Failed");
REQUIRE(false);
}

/* GpuId[0] for gfx90a exists--> 1 for yes and 0 for no */
int GpuId[1] = {0};
p = fork();

if (p < 0) {
fprintf(stderr, "fork Failed");
REQUIRE(false);
} else if (p > 0) { // parent process
close(fd1[1]); // Close writing end of first pipe
// Wait for child to send a string
wait(NULL);
// Read string from child and close reading end.
read(fd1[0], GpuId, 2 * sizeof(int));
close(fd1[0]);
if (GpuId[0] == 0) {
WARN("This test is applicable for MI200."
"Skipping the test!!");
exit(0);
}
} else { // child process
close(fd1[0]); // Close read end of first pipe
hipDeviceProp_t prop;
HIPCHECK(hipGetDeviceProperties(&prop, 0));
char *p = NULL;
p = strstr(prop.gcnArchName, "gfx90a");
if (p) {
WARN("gfx90a gpu found on this system!!");
GpuId[0] = 1;
}
// Write concatenated string and close writing end
write(fd1[1], GpuId, 2 * sizeof(int));
close(fd1[1]);
exit(0);
}
int stat = 0;
if (fork() == 0) {
hipDeviceProp_t prop;
HIPCHECK(hipGetDeviceProperties(&prop, 0));
char *p = NULL;
p = strstr(prop.gcnArchName, "xnack+");
if (p) {
int stat = 0;
int managed = 0;
HIP_CHECK(hipDeviceGetAttribute(&managed, hipDeviceAttributeManagedMemory,
0));
Expand All @@ -275,25 +177,10 @@ TEST_CASE("Unit_malloc_CoherentTstWthAdvise") {
SquareKrnl<<<1, 1, 0, strm>>>(Ptr);
HIP_CHECK(hipStreamSynchronize(strm));
HIP_CHECK(hipStreamDestroy(strm));
if (*Ptr == 16) {
// exit() with code 10 which indicates pass
free(Ptr);
exit(10);
} else {
// exit() with code 9 which indicates fail
free(Ptr);
exit(9);
}
} else {
SUCCEED("GPU 0 doesn't support hipDeviceAttributeManagedMemory "
"attribute. Hence skipping the testing with Pass result.\n");
REQUIRE (*Ptr == 16);
}
} else {
wait(&stat);
int Result = WEXITSTATUS(stat);
if (Result != 10) {
REQUIRE(false);
}
HipTest::HIP_SKIP_TEST("GPU is not xnack enabled hence skipping the test...\n");
}
}
#endif
Expand All @@ -303,55 +190,12 @@ TEST_CASE("Unit_malloc_CoherentTstWthAdvise") {
// The following test is failing on Nvidia platform hence disabling it for now
#if HT_AMD
TEST_CASE("Unit_mmap_CoherentTst") {
if ((setenv("HSA_XNACK", "1", 1)) != 0) {
WARN("Unable to turn on HSA_XNACK, hence terminating the Test case!");
REQUIRE(false);
}
// The following code block is used to check for gfx906/8 so as to skip if
// any of the gpus available
int fd1[2]; // Used to store two ends of first pipe
pid_t p;
if (pipe(fd1) == -1) {
fprintf(stderr, "Pipe Failed");
REQUIRE(false);
}

/* GpuId[0] for gfx90a exists--> 1 for yes and 0 for no */
int GpuId[1] = {0};
p = fork();

if (p < 0) {
fprintf(stderr, "fork Failed");
REQUIRE(false);
} else if (p > 0) { // parent process
close(fd1[1]); // Close writing end of first pipe
// Wait for child to send a string
wait(NULL);
// Read string from child and close reading end.
read(fd1[0], GpuId, 2 * sizeof(int));
close(fd1[0]);
if (GpuId[0] == 0) {
WARN("This test is not applicable for MI200."
"Skipping the test!!");
exit(0);
}
} else { // child process
close(fd1[0]); // Close read end of first pipe
hipDeviceProp_t prop;
HIPCHECK(hipGetDeviceProperties(&prop, 0));
char *p = NULL;
p = strstr(prop.gcnArchName, "gfx90a");
if (p) {
WARN("gfx90a gpu found on this system!!");
GpuId[0] = 1;
}
// Write concatenated string and close writing end
write(fd1[1], GpuId, 2 * sizeof(int));
close(fd1[1]);
exit(0);
}
int stat = 0;
if (fork() == 0) {
hipDeviceProp_t prop;
HIPCHECK(hipGetDeviceProperties(&prop, 0));
char *p = NULL;
p = strstr(prop.gcnArchName, "xnack+");
if (p) {
int stat = 0;
int managed = 0;
HIP_CHECK(hipDeviceGetAttribute(&managed, hipDeviceAttributeManagedMemory,
0));
Expand All @@ -371,21 +215,10 @@ TEST_CASE("Unit_mmap_CoherentTst") {
if (err != 0) {
WARN("munmap failed\n");
}
if (YES_COHERENT) {
exit(10);
} else {
exit(9);
}
} else {
SUCCEED("GPU 0 doesn't support hipDeviceAttributeManagedMemory "
"attribute. Hence skipping the testing with Pass result.\n");
}
REQUIRE(YES_COHERENT);
}
} else {
wait(&stat);
int Result = WEXITSTATUS(stat);
if (Result != 10) {
REQUIRE(false);
}
HipTest::HIP_SKIP_TEST("GPU is not xnack enabled hence skipping the test...\n");
}
}
#endif
Expand All @@ -395,55 +228,12 @@ TEST_CASE("Unit_mmap_CoherentTst") {
// The following test is failing on Nvidia platform hence disabling it for now
#if HT_AMD
TEST_CASE("Unit_mmap_CoherentTstWthAdvise") {
if ((setenv("HSA_XNACK", "1", 1)) != 0) {
WARN("Unable to turn on HSA_XNACK, hence terminating the Test case!");
REQUIRE(false);
}
// The following code block is used to check for gfx906/8 so as to skip if
// any of the gpus available
int fd1[2]; // Used to store two ends of first pipe
pid_t p;
if (pipe(fd1) == -1) {
fprintf(stderr, "Pipe Failed");
REQUIRE(false);
}

/* GpuId[0] for gfx90a exists--> 1 for yes and 0 for no */
int GpuId[1] = {0};
p = fork();

if (p < 0) {
fprintf(stderr, "fork Failed");
REQUIRE(false);
} else if (p > 0) { // parent process
close(fd1[1]); // Close writing end of first pipe
// Wait for child to send a string
wait(NULL);
// Read string from child and close reading end.
read(fd1[0], GpuId, 2 * sizeof(int));
close(fd1[0]);
if (GpuId[0] == 0) {
WARN("This test is applicable for MI200."
"Skipping the test!!");
exit(0);
}
} else { // child process
close(fd1[0]); // Close read end of first pipe
hipDeviceProp_t prop;
HIPCHECK(hipGetDeviceProperties(&prop, 0));
char *p = NULL;
p = strstr(prop.gcnArchName, "gfx90a");
if (p) {
WARN("gfx90a gpu found on this system!!");
GpuId[0] = 1;
}
// Write concatenated string and close writing end
write(fd1[1], GpuId, 2 * sizeof(int));
close(fd1[1]);
exit(0);
}
int stat = 0;
if (fork() == 0) {
hipDeviceProp_t prop;
HIPCHECK(hipGetDeviceProperties(&prop, 0));
char *p = NULL;
p = strstr(prop.gcnArchName, "xnack+");
if (p) {
int stat = 0;
int managed = 0;
HIP_CHECK(hipDeviceGetAttribute(&managed, hipDeviceAttributeManagedMemory,
0));
Expand All @@ -466,26 +256,15 @@ TEST_CASE("Unit_mmap_CoherentTstWthAdvise") {
bool IfTstPassed = false;
if (*Ptr == 81) {
IfTstPassed = true;
}
}
int err = munmap(Ptr, SIZE);
if (err != 0) {
WARN("munmap failed\n");
}
if (IfTstPassed) {
exit(10);
} else {
exit(9);
}
} else {
SUCCEED("GPU 0 doesn't support hipDeviceAttributeManagedMemory "
"attribute. Hence skipping the testing with Pass result.\n");
}
REQUIRE(IfTstPassed);
}
} else {
wait(&stat);
int Result = WEXITSTATUS(stat);
if (Result != 10) {
REQUIRE(false);
}
HipTest::HIP_SKIP_TEST("GPU is not xnack enabled hence skipping the test...\n");
}
}
#endif
Expand Down
3 changes: 2 additions & 1 deletion catch/unit/memory/CMakeLists.txt
View file
Original file line number Diff line number Diff line change
Expand Up @@ -108,7 +108,8 @@ set(TEST_SRC
hipMemRangeGetAttributes.cc
hipStreamAttachMemAsync.cc
hipMemRangeGetAttributes_old.cc
hipMemGetAddressRange.cc)
hipMemGetAddressRange.cc
hipHmmOvrSubscriptionTst.cc)

if(HIP_PLATFORM MATCHES "amd")
set(TEST_SRC
Expand Down
Loading

0 comments on commit 9cf8b32

Please sign in to comment.