AI/rocm-systems
2
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

SWDEV-465073 - Update hipClassKernel_Virtual test

Browse Source 
Compiler has changed to match the device side struct layout with
the host side and hence the test is updated to verify if struct sizes
of both host and device are equal instead of checking for hardcoded
struct sizes

Change-Id: Idd9ea1f04b413b7aa30f03555fd8e0ee11cd0f75


[ROCm/hip-tests commit: 5f1dfc3851]
This commit is contained in:
Satyanvesh Dittakavi
2024-06-06 16:47:23 +00:00
parent 35782c04bc
commit 4a0e156ae2
2 changed files with 25 additions and 9 deletions
Download Patch File Download Diff File Expand all files Collapse all files
projects/hip-tests/catch/unit/compiler/hipClassKernel.cc
+16 -8
View File
@@ -164,13 +164,13 @@ TEST_CASE("Unit_hipClassKernel_Size") {
}
__global__ void
sizeVirtualClassKernel(bool* result_ecd) {
sizeVirtualClassKernel(bool* result_ecd, refStructSizes structSizes) {
int tid = threadIdx.x + blockIdx.x * blockDim.x;
result_ecd[tid] = (sizeof(testSizeDV) == 16)
&& (sizeof(testSizeDerivedDV) == 16)
&& (sizeof(testSizeVirtDerPack) == 24)
&& (sizeof(testSizeVirtDer) == 24)
&& (sizeof(testSizeDerMulti) == 48) ;
result_ecd[tid] = (structSizes.sizeOftestSizeDV == sizeof(testSizeDV))
&& (structSizes.sizeOftestSizeDerivedDV == sizeof(testSizeDerivedDV))
&& (structSizes.sizeOftestSizeVirtDer = sizeof(testSizeVirtDer))
&& (structSizes.sizeOftestSizeVirtDerPack = sizeof(testSizeVirtDerPack))
&& (structSizes.sizeOftestSizeDerMulti = sizeof(testSizeDerMulti));
}
TEST_CASE("Unit_hipClassKernel_Virtual") {
@@ -178,12 +178,20 @@ TEST_CASE("Unit_hipClassKernel_Virtual") {
result_ech = AllocateHostMemory();
result_ecd = AllocateDeviceMemory();
struct refStructSizes structSizes;
structSizes.sizeOftestSizeDV = sizeof(testSizeDV);
structSizes.sizeOftestSizeDerivedDV = sizeof(testSizeDerivedDV);
structSizes.sizeOftestSizeVirtDer = sizeof(testSizeVirtDer);
structSizes.sizeOftestSizeVirtDerPack = sizeof(testSizeVirtDerPack);
structSizes.sizeOftestSizeDerMulti = sizeof(testSizeDerMulti);
hipLaunchKernelGGL(sizeVirtualClassKernel,
dim3(BLOCKS),
dim3(THREADS_PER_BLOCK),
0,
0,
result_ecd);
result_ecd,
structSizes);
VerifyResult(result_ech,result_ecd);
FreeMem(result_ech,result_ecd);
@@ -215,4 +223,4 @@ TEST_CASE("Unit_hipClassKernel_Value") {
VerifyResult(result_ech,result_ecd);
FreeMem(result_ech,result_ecd);
}
}
projects/hip-tests/catch/unit/compiler/hipClassKernel.h
+9 -1
View File
@@ -152,6 +152,14 @@ class testSizeDerivedDV : testSizeDV {
int iDDV;
};
struct refStructSizes {
size_t sizeOftestSizeDV;
size_t sizeOftestSizeDerivedDV;
size_t sizeOftestSizeVirtDer;
size_t sizeOftestSizeVirtDerPack;
size_t sizeOftestSizeDerMulti;
};
#pragma pack(push, 1)
class testSizeP1 {
public:
@@ -231,4 +239,4 @@ void FreeMem(bool* result_ech, bool* result_ecd){
HIPCHECK(hipFree(result_ecd));
}
#endif // _HIPCLASSKERNEL_H_
#endif // _HIPCLASSKERNEL_H_