diff --git a/tests/src/CMakeLists.txt b/tests/src/CMakeLists.txt index b4627c9ed4..f8408df4ff 100644 --- a/tests/src/CMakeLists.txt +++ b/tests/src/CMakeLists.txt @@ -130,6 +130,8 @@ make_hip_executable (hipHostAlloc hipHostAlloc.cpp) make_hip_executable (hipStreamL5 hipStreamL5.cpp) make_hip_executable (hipHostGetFlags hipHostGetFlags.cpp) make_hip_executable (hipHostRegister hipHostRegister.cpp) +make_hip_executable (hipRandomMemcpyAsync hipRandomMemcpyAsync.cpp) + target_link_libraries(hipMathFunctionsHost m) make_test(hip_ballot " " ) @@ -155,5 +157,5 @@ make_test(hipHostGetFlags " ") make_test(hipHcc " " ) make_test(hipHostRegister " ") make_test(hipStreamL5 " ") - +make_test(hipRandomMemcpyAsync " ") make_hipify_test(specialFunc.cu ) diff --git a/tests/src/hipRandomMemcpyAsync.cpp b/tests/src/hipRandomMemcpyAsync.cpp new file mode 100644 index 0000000000..cc4a5d725a --- /dev/null +++ b/tests/src/hipRandomMemcpyAsync.cpp @@ -0,0 +1,80 @@ +#include +#include +#include +#include "hip_runtime.h" +#include "test_common.h" + +#define WIDTH 1024 +#define HEIGHT 1024 + +#define NUM (WIDTH*HEIGHT) + +#define THREADS_PER_BLOCK_X 16 +#define THREADS_PER_BLOCK_Y 16 +#define THREADS_PER_BLOCK_Z 1 + +int main() { + + int *hostA; + int *hostB; + + int *deviceA; + int *deviceB; + + int i; + int errors; + + hostA = (int *)malloc(NUM * sizeof(int)); + hostB = (int *)malloc(NUM * sizeof(int)); + + // initialize the input data + for (i = 0; i < NUM; i++) { + hostB[i] = i; + } + + HIPCHECK(hipMalloc((void**)&deviceA, NUM * sizeof(int))); + HIPCHECK(hipMalloc((void**)&deviceB, NUM * sizeof(int))); + + hipStream_t s; + HIPCHECK(hipStreamCreate(&s)); + + + // hostB -> deviceB -> hostA +#define ASYNC 1 +#if ASYNC + HIPCHECK(hipMemcpyAsync(deviceB, hostB, NUM*sizeof(int), hipMemcpyHostToDevice, s)); + HIPCHECK(hipMemcpyAsync(hostA, deviceB, NUM*sizeof(int), hipMemcpyDeviceToHost, s)); +#else + HIPCHECK(hipMemcpy(deviceB, hostB, NUM*sizeof(int), hipMemcpyHostToDevice)); + HIPCHECK(hipMemcpy(hostA, deviceB, NUM*sizeof(int), hipMemcpyDeviceToHost)); +#endif + + HIPCHECK(hipStreamSynchronize(s)); + HIPCHECK(hipDeviceSynchronize()); + + // verify the results + errors = 0; + for (i = 0; i < NUM; i++) { + if (hostA[i] != (hostB[i])) { + errors++; + } + } + + HIPCHECK(hipStreamDestroy(s)); + + HIPCHECK(hipFree(deviceA)); + HIPCHECK(hipFree(deviceB)); + + free(hostA); + free(hostB); + + //hipResetDefaultAccelerator(); + + if(errors != 0){ + HIPASSERT(1 == 2); + }else{ + passed(); + } + + return errors; +}