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

Refactor module API test.

Browse Source 
- Add PASSED/FAIL indication.
- Set args using struct rather than void* array.

Change-Id: Ic924f88c49cc46979b12b7fef8650081e3b5f58c
This commit is contained in:
Ben Sander
2016-10-14 23:19:25 -05:00
parent 84283d0801
commit 3ae3c39e44
1 changed files with 35 additions and 20 deletions
Download Patch File Download Diff File Expand all files Collapse all files
samples/0_Intro/module_api/runKernel.cpp
+35 -20
View File
@@ -60,32 +60,37 @@ int main(){
uint32_t len = LEN;
uint32_t one = 1;
std::vector<void*>argBuffer(5);
uint32_t *ptr32_t = (uint32_t*)&argBuffer[0];
memcpy(ptr32_t + 0, &one, sizeof(uint32_t));
memcpy(ptr32_t + 1, &one, sizeof(uint32_t));
memcpy(ptr32_t + 2, &one, sizeof(uint32_t));
memcpy(ptr32_t + 3, &len, sizeof(uint32_t));
memcpy(ptr32_t + 4, &one, sizeof(uint32_t));
memcpy(ptr32_t + 5, &one, sizeof(uint32_t));
memcpy(&argBuffer[3], &Ad, sizeof(void*));
memcpy(&argBuffer[4], &Bd, sizeof(void*));
struct {
uint32_t _hidden[6];
void * _Ad;
void * _Bd;
} args;
for (int i=0; i<6; i++) {
args._hidden[i] = 0;
}
args._Ad = Ad;
args._Bd = Bd;
#endif
#ifdef __HIP_PLATFORM_NVCC__
uint32_t one = 1;
std::vector<void*>argBuffer(3);
uint32_t *ptr32_t = (uint32_t*)&argBuffer[0];
memcpy(ptr32_t + 0, &one, sizeof(uint32_t));
memcpy(&argBuffer[1], &Ad, sizeof(void*));
memcpy(&argBuffer[2], &Bd, sizeof(void*));
struct {
uint32_t _hidden[1];
void * _Ad;
void * _Bd;
} args;
args._hidden[0] = 0;
args._Ad = Ad;
args._Bd = Bd;
#endif
size_t size = argBuffer.size()*sizeof(void*);
size_t size = sizeof(args);
void *config[] = {
HIP_LAUNCH_PARAM_BUFFER_POINTER, &argBuffer[0],
HIP_LAUNCH_PARAM_BUFFER_POINTER, &args,
HIP_LAUNCH_PARAM_BUFFER_SIZE, &size,
HIP_LAUNCH_PARAM_END
};
@@ -93,10 +98,20 @@ int main(){
hipModuleLaunchKernel(Function, 1, 1, 1, LEN, 1, 1, 0, 0, NULL, (void**)&config);
hipMemcpyDtoH(B, Bd, SIZE);
for(uint32_t i=LEN-4;i<LEN;i++){
std::cout<<A[i]<<" - "<<B[i]<<std::endl;
int mismatchCount = 0;
for(uint32_t i=0;i<LEN;i++){
if (A[i] != B[i]) {
mismatchCount++;
std::cout<<"error: mismatch " << A[i]<<" != "<<B[i]<<std::endl;
}
}
if (mismatchCount == 0) {
std::cout << "PASSED!\n";
} else {
std::cout << "FAILED!\n";
};
hipCtxDestroy(context);
return 0;
}