Fix sample to use kernelargs for launch
This commit is contained in:
@@ -21,7 +21,6 @@ THE SOFTWARE.
|
||||
*/
|
||||
|
||||
#include "hip/hip_runtime.h"
|
||||
#include "hip/hip_runtime_api.h"
|
||||
#include <iostream>
|
||||
#include <fstream>
|
||||
#include <vector>
|
||||
@@ -29,20 +28,18 @@ THE SOFTWARE.
|
||||
#define LEN 64
|
||||
#define SIZE LEN << 2
|
||||
|
||||
#define fileName "test.co"
|
||||
#define kernel_name "vadd"
|
||||
#define fileName "vcpy_kernel.code"
|
||||
#define kernel_name "hello_world"
|
||||
|
||||
int main() {
|
||||
float *A, *B, *C;
|
||||
hipDeviceptr_t Ad, Bd, Cd;
|
||||
float *A, *B;
|
||||
hipDeviceptr_t Ad, Bd;
|
||||
A = new float[LEN];
|
||||
B = new float[LEN];
|
||||
C = new float[LEN];
|
||||
|
||||
for (uint32_t i = 0; i < LEN; i++) {
|
||||
A[i] = i * 1.0f;
|
||||
B[i] = 1.0f;
|
||||
C[i] = 0.0f;
|
||||
B[i] = 0.0f;
|
||||
}
|
||||
|
||||
hipInit(0);
|
||||
@@ -53,28 +50,25 @@ int main() {
|
||||
|
||||
hipMalloc((void**)&Ad, SIZE);
|
||||
hipMalloc((void**)&Bd, SIZE);
|
||||
hipMalloc((void**)&Cd, SIZE);
|
||||
|
||||
hipMemcpyHtoD(Ad, A, SIZE);
|
||||
hipMemcpyHtoD(Bd, B, SIZE);
|
||||
hipMemcpyHtoD(Cd, C, SIZE);
|
||||
|
||||
hipModule_t Module;
|
||||
hipFunction_t Function;
|
||||
hipModuleLoad(&Module, fileName);
|
||||
hipModuleGetFunction(&Function, Module, kernel_name);
|
||||
|
||||
int n = LEN;
|
||||
void* args[4] = {&Ad, &Bd, &Cd, &n};
|
||||
void* args[2] = {&Ad, &Bd};
|
||||
|
||||
hipModuleLaunchKernel(Function, 1, 1, 1, LEN, 1, 1, 0, 0, args, nullptr);
|
||||
|
||||
hipMemcpyDtoH(C, Cd, SIZE);
|
||||
hipMemcpyDtoH(B, Bd, SIZE);
|
||||
int mismatchCount = 0;
|
||||
for (uint32_t i = 0; i < LEN; i++) {
|
||||
if (A[i] + B[i] != C[i]) {
|
||||
if (A[i] != B[i]) {
|
||||
mismatchCount++;
|
||||
std::cout << "error: mismatch " << A[i] << " + " << B[i] << " != " << C[i] << std::endl;
|
||||
std::cout << "error: mismatch " << A[i] << " != " << C[i] << std::endl;
|
||||
}
|
||||
}
|
||||
|
||||
@@ -84,6 +78,10 @@ int main() {
|
||||
std::cout << "FAILED!\n";
|
||||
};
|
||||
|
||||
hipFree(Ad);
|
||||
hipFree(Bd);
|
||||
delete A;
|
||||
delete B;
|
||||
hipCtxDestroy(context);
|
||||
return 0;
|
||||
}
|
||||
|
||||
Reference in New Issue
Block a user