diff --git a/hipamd/api/hip/hip_hcc.def.in b/hipamd/api/hip/hip_hcc.def.in index 70eaa66870..45aa1503cc 100644 --- a/hipamd/api/hip/hip_hcc.def.in +++ b/hipamd/api/hip/hip_hcc.def.in @@ -60,6 +60,7 @@ hipGetDeviceProperties hipGetErrorName hipGetErrorString hipGetLastError +hipMemAllocHost hipHostAlloc hipHostFree hipHostGetDevicePointer @@ -77,6 +78,7 @@ hipMalloc3DArray hipArrayCreate hipArray3DCreate hipMallocArray +hipMemAllocPitch hipMallocPitch hipMemcpy hipMemcpyParam2D @@ -105,9 +107,12 @@ hipGetSymbolSize hipMemGetInfo hipMemPtrGetInfo hipMemset +hipMemsetD16 hipMemsetD32 hipMemset2D hipMemsetAsync +hipMemsetD8Async +hipMemsetD16Async hipMemsetD32Async hipMemset2DAsync hipMemsetD8 diff --git a/hipamd/api/hip/hip_hcc.map.in b/hipamd/api/hip/hip_hcc.map.in index 217164cafb..8535ab72f1 100644 --- a/hipamd/api/hip/hip_hcc.map.in +++ b/hipamd/api/hip/hip_hcc.map.in @@ -61,6 +61,7 @@ global: hipGetErrorName; hipGetErrorString; hipGetLastError; + hipMemAllocHost; hipHostAlloc; hipHostFree; hipHostGetDevicePointer; @@ -79,6 +80,7 @@ global: hipArray3DCreate; hipMallocArray; hipMallocPitch; + hipMemAllocPitch; hipMemcpy; hipMemcpyParam2D; hipMemcpy2D; @@ -106,9 +108,12 @@ global: hipMemGetInfo; hipMemPtrGetInfo; hipMemset; + hipMemsetD16; hipMemsetD32; hipMemset2D; hipMemsetAsync; + hipMemsetD8Async; + hipMemsetD16Async; hipMemsetD32Async; hipMemset2DAsync; hipMemsetD8; diff --git a/hipamd/api/hip/hip_memory.cpp b/hipamd/api/hip/hip_memory.cpp index 5abfb7e365..174bd14e5b 100644 --- a/hipamd/api/hip/hip_memory.cpp +++ b/hipamd/api/hip/hip_memory.cpp @@ -269,6 +269,14 @@ hipError_t hipMemsetAsync(void* dst, int value, size_t sizeBytes, hipStream_t st HIP_RETURN(ihipMemset(dst, value, sizeof(char), sizeBytes, stream, true)); } +hipError_t hipMemsetD16Async(hipDeviceptr_t dst, unsigned short value, size_t count, hipStream_t stream){ + HIP_INIT_API(hipMemsetD16Async, dst, value, count, stream); + + assert(0 && "Unimplemented"); + + HIP_RETURN(hipErrorUnknown); +} + hipError_t hipMemsetD32Async(hipDeviceptr_t dst, int value, size_t count, hipStream_t stream) { HIP_INIT_API(hipMemsetD32Async, dst, value, count, stream); @@ -281,6 +289,14 @@ hipError_t hipMemset(void* dst, int value, size_t sizeBytes) { HIP_RETURN(ihipMemset(dst, value, sizeof(char), sizeBytes, nullptr)); } +hipError_t hipMemsetD16(hipDeviceptr_t dst, unsigned short value, size_t count){ + HIP_INIT_API(hipMemsetD16, dst, value, count); + + assert(0 && "Unimplemented"); + + HIP_RETURN(hipErrorUnknown); +} + hipError_t hipMemsetD32(hipDeviceptr_t dst, int value, size_t count) { HIP_INIT_API(hipMemsetD32, dst, value, count); @@ -1341,16 +1357,14 @@ hipError_t hipMemset2DAsync(void* dst, size_t pitch, int value, HIP_RETURN(ihipMemset2D(dst, pitch, value, width, height, *queue, true)); } -hipError_t hipMemsetD8(hipDeviceptr_t dst, unsigned char value, size_t sizeBytes) { - HIP_INIT_API(hipMemsetD8, dst, value, sizeBytes); +hipError_t ihipMemsetD8(hipDeviceptr_t dst, unsigned char value, size_t sizeBytes, hipStream_t stream) { if (dst == nullptr) { - HIP_RETURN(hipErrorInvalidValue); + return hipErrorInvalidValue; } - hip::syncStreams(); - amd::HostQueue* queue = hip::getNullStream(); size_t offset = 0; + amd::HostQueue* queue = hip::getQueue(stream); amd::Command::EventWaitList waitList; amd::Memory* memory = getMemoryObject(dst, offset); if (memory != nullptr) { @@ -1373,7 +1387,33 @@ hipError_t hipMemsetD8(hipDeviceptr_t dst, unsigned char value, size_t sizeBytes memset(dst, value, sizeBytes); } - HIP_RETURN(hipSuccess); + return hipSuccess; +} + +hipError_t hipMemsetD8(hipDeviceptr_t dst, unsigned char value, size_t sizeBytes) { + HIP_INIT_API(hipMemsetD8, dst, value, sizeBytes); + + HIP_RETURN(ihipMemsetD8(dst, value, sizeBytes, nullptr)); +} + +hipError_t hipMemsetD8Async(hipDeviceptr_t dst, unsigned char value, size_t sizeBytes, + hipStream_t stream) { + HIP_INIT_API(hipMemsetD8Async, dst, value, sizeBytes, stream); + + HIP_RETURN(ihipMemsetD8(dst, value, sizeBytes, stream)); +} + +hipError_t hipMemAllocPitch(hipDeviceptr_t* dptr, size_t* pitch, size_t widthInBytes, + size_t height, unsigned int elementSizeBytes) { + HIP_INIT_API(hipMemAllocPitch, dptr, pitch, widthInBytes, height, elementSizeBytes); + + HIP_RETURN(hipMallocPitch(dptr, pitch, widthInBytes, height)); +} + +hipError_t hipMemAllocHost(void** ptr, size_t size) { + HIP_INIT_API(hipMemAllocHost, ptr, size); + + HIP_RETURN(hipHostMalloc(ptr, size, 0)); } hipError_t hipIpcGetMemHandle(hipIpcMemHandle_t* handle, void* dev_ptr) {