From 2c8d0de12daae4c68ed35a45b16dbb442cafa300 Mon Sep 17 00:00:00 2001 From: foreman Date: Mon, 1 Apr 2019 11:51:06 -0400 Subject: [PATCH] P4 to Git Change 1764069 by michliao@hliao-dev-00-hip-workspace on 2019/04/01 11:23:34 SWDEV-144570 - Handle zero-byte memset & memcpy - Properly handle zero-byte memset & memcpy by skipping the real stuff. RB: http://ocltc.amd.com/reviews/r/17062/ Affected files ... ... //depot/stg/opencl/drivers/opencl/api/hip/hip_memory.cpp#48 edit [ROCm/hip commit: b392004e71d0d402870b79ce5c70cdaf8def3037] --- projects/hip/api/hip/hip_memory.cpp | 9 +++++++++ 1 file changed, 9 insertions(+) diff --git a/projects/hip/api/hip/hip_memory.cpp b/projects/hip/api/hip/hip_memory.cpp index f20b26aaa0..ae052d6288 100644 --- a/projects/hip/api/hip/hip_memory.cpp +++ b/projects/hip/api/hip/hip_memory.cpp @@ -76,6 +76,11 @@ hipError_t ihipMalloc(void** ptr, size_t sizeBytes, unsigned int flags) hipError_t ihipMemcpy(void* dst, const void* src, size_t sizeBytes, hipMemcpyKind kind, amd::HostQueue& queue, bool isAsync = false) { + if (sizeBytes == 0) { + // Skip if nothing needs writing. + return hipSuccess; + } + amd::Command* command = nullptr; amd::Command::EventWaitList waitList; @@ -114,6 +119,10 @@ hipError_t ihipMemcpy(void* dst, const void* src, size_t sizeBytes, hipMemcpyKin hipError_t ihipMemset(void* dst, int value, size_t valueSize, size_t sizeBytes, hipStream_t stream, bool isAsync = false) { + if (sizeBytes == 0) { + // Skip if nothing needs filling. + return hipSuccess; + } if (dst == nullptr) { return hipErrorInvalidValue;