From 1a27e5134ea18463d369ea455becb386dc2ffd89 Mon Sep 17 00:00:00 2001 From: Ben Sander Date: Mon, 14 Mar 2016 14:40:41 -0500 Subject: [PATCH] enable DB, comments --- include/hcc_detail/hip_runtime_api.h | 3 ++- src/hip_hcc.cpp | 4 +++- 2 files changed, 5 insertions(+), 2 deletions(-) diff --git a/include/hcc_detail/hip_runtime_api.h b/include/hcc_detail/hip_runtime_api.h index 867996042e..44dffc8839 100644 --- a/include/hcc_detail/hip_runtime_api.h +++ b/include/hcc_detail/hip_runtime_api.h @@ -720,7 +720,8 @@ hipError_t hipMemcpyToSymbol(const char* symbolName, const void *src, size_t siz /** * @brief Copy data from src to dst asynchronously. * - * TODO: cudaErrorInvalidMemcpyDirection error code is not supported right now, use hipErrorUnknown for now + * @warning If host or dest are not pinned, the memory copy will be performed synchronously. For best performance, use hipHostAlloc to + * allocate host memory that is transferred asynchronously. * * @param[out] dst Data being copy to * @param[in] src Data being copy from diff --git a/src/hip_hcc.cpp b/src/hip_hcc.cpp index 5bb0d8a404..d2fae41667 100644 --- a/src/hip_hcc.cpp +++ b/src/hip_hcc.cpp @@ -91,7 +91,7 @@ int HIP_ONESHOT_COPY_DEP = 1; // TODO - setting this =1 is a good thing, r // Compile debug trace mode - this prints debug messages to stderr when env var HIP_DB is set. // May be set to 0 to remove debug if checks - possible code size and performance difference? -#define COMPILE_DB_TRACE 0 +#define COMPILE_DB_TRACE 1 // Color defs for debug messages: @@ -1053,6 +1053,7 @@ void ihipInit() } + INLINE bool ihipIsValidDevice(unsigned deviceIndex) { // deviceIndex is unsigned so always > 0 @@ -1678,6 +1679,7 @@ hipError_t hipEventRecord(hipEvent_t event, hipStream_t stream) // If stream == NULL, wait on all queues. // This matches behavior described in CUDA 7 RT APIs, which say that "This function uses standard default stream semantics". // TODO-HCC fix this - is CUDA this conservative or still uses device timestamps? + // TODO-HCC can we use barrier or event marker to implement better solution? ihipDevice_t *device = ihipGetTlsDefaultDevice(); ihipWaitNullStream(device);