diff --git a/hipamd/samples/0_Intro/module_api/Makefile b/hipamd/samples/0_Intro/module_api/Makefile index 270d4c1211..8d0237826f 100644 --- a/hipamd/samples/0_Intro/module_api/Makefile +++ b/hipamd/samples/0_Intro/module_api/Makefile @@ -5,7 +5,7 @@ endif HIPCC=$(HIP_PATH)/bin/hipcc HIP_PLATFORM=$(shell $(HIP_PATH)/bin/hipconfig --compiler) -all: vcpy_kernel.code runKernel.hip.out launchKernelHcc.hip.out +all: vcpy_kernel.code runKernel.hip.out launchKernelHcc.hip.out defaultDriver.hip.out runKernel.hip.out: runKernel.cpp $(HIPCC) $(HIPCC_FLAGS) $< -o $@ @@ -13,8 +13,8 @@ runKernel.hip.out: runKernel.cpp launchKernelHcc.hip.out: launchKernelHcc.cpp $(HIPCC) $(HIPCC_FLAGS) $< -o $@ -#defaultDriver.hip.out: defaultDriver.cpp -# $(HIPCC) $(HIPCC_FLAGS) $< -o $@ +defaultDriver.hip.out: defaultDriver.cpp + $(HIPCC) $(HIPCC_FLAGS) $< -o $@ vcpy_kernel.code: vcpy_kernel.cpp $(HIPCC) --genco $(GENCO_FLAGS) $^ -o $@ diff --git a/hipamd/samples/0_Intro/module_api/defaultDriver.cpp b/hipamd/samples/0_Intro/module_api/defaultDriver.cpp index b6abfcc3a9..9fa0b0ee09 100644 --- a/hipamd/samples/0_Intro/module_api/defaultDriver.cpp +++ b/hipamd/samples/0_Intro/module_api/defaultDriver.cpp @@ -21,7 +21,6 @@ THE SOFTWARE. */ #include "hip/hip_runtime.h" -#include "hip/hip_runtime_api.h" #include #include #include @@ -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; }