From 7912e615022de55874fbd09305fcf4e355ffa182 Mon Sep 17 00:00:00 2001 From: Ben Sander Date: Fri, 23 Jun 2017 10:38:29 -0500 Subject: [PATCH] Clean up old USE_* and RELEASE.md notes. --- RELEASE.md | 6 ------ include/hip/hcc_detail/host_defines.h | 5 ----- src/hip_hcc.cpp | 3 --- tests/src/hipPointerAttrib.cpp | 9 --------- tests/src/runtimeApi/memory/p2p_copy_coherency.cpp | 4 ---- 5 files changed, 27 deletions(-) diff --git a/RELEASE.md b/RELEASE.md index 5787c59881..d6f3ec594c 100644 --- a/RELEASE.md +++ b/RELEASE.md @@ -2,12 +2,6 @@ We have attempted to document known bugs and limitations - in particular the [HIP Kernel Language](docs/markdown/hip_kernel_language.md) document uses the phrase "Under Development", and the [HIP Runtime API bug list](http://gpuopen-professionalcompute-tools.github.io/HIP/bug.html) lists known bugs. -Upcoming: -- Stability: Enforce periodic host synchronization to reclaim resources if the application has launched a large - number of commands (>1K) without synchronizing. -- Register keyword now silently ignored on HCC (previously would emit warning). -- Doc updates: Add some more frequently asked questions to FAQ, fix TOC in some files, review. -- Cookbook. =================================================================================================== diff --git a/include/hip/hcc_detail/host_defines.h b/include/hip/hcc_detail/host_defines.h index 140cbb0678..212fd650a3 100644 --- a/include/hip/hcc_detail/host_defines.h +++ b/include/hip/hcc_detail/host_defines.h @@ -28,7 +28,6 @@ THE SOFTWARE. #ifndef HIP_INCLUDE_HIP_HCC_DETAIL_HOST_DEFINES_H #define HIP_INCLUDE_HIP_HCC_DETAIL_HOST_DEFINES_H -#define USE_PROMOTE_FREE_HCC 1 // Add guard to Generic Grid Launch method #ifndef GENERIC_GRID_LAUNCH @@ -61,11 +60,7 @@ THE SOFTWARE. */ // _restrict is supported by the compiler #define __shared__ tile_static -#if USE_PROMOTE_FREE_HCC==1 #define __constant__ __attribute__((hc)) -#else -#define __constant__ ADDRESS_SPACE_1 -#endif #else // Non-HCC compiler diff --git a/src/hip_hcc.cpp b/src/hip_hcc.cpp index d826a0cec3..061714070e 100644 --- a/src/hip_hcc.cpp +++ b/src/hip_hcc.cpp @@ -48,9 +48,6 @@ THE SOFTWARE. #include "env.h" -// needs HCC change for hc::no_scope -#define USE_NO_SCOPE 1 - //================================================================================================= //Global variables: //================================================================================================= diff --git a/tests/src/hipPointerAttrib.cpp b/tests/src/hipPointerAttrib.cpp index 7a2ab64bea..bddbff5ce0 100644 --- a/tests/src/hipPointerAttrib.cpp +++ b/tests/src/hipPointerAttrib.cpp @@ -32,7 +32,6 @@ THE SOFTWARE. #endif -#define USE_AV_COPY (__hcc_workweek__ >= 16351) size_t Nbytes = 0; @@ -410,21 +409,13 @@ void thread_noise_generator(int iters, size_t numBuffers, Dir addDir, Dir remove if (addDir == Up) { for (char *p = basePtr; p=0; p-=bufferSize) { -#if USE_AV_COPY hc::AmPointerInfo info(p, p, bufferSize, acc, false, false); hc::am_memtracker_add(p, info); -#else - hc::am_memtracker_add(p, bufferSize, acc, false); -#endif } } diff --git a/tests/src/runtimeApi/memory/p2p_copy_coherency.cpp b/tests/src/runtimeApi/memory/p2p_copy_coherency.cpp index a5d79464d0..9fadebea1e 100644 --- a/tests/src/runtimeApi/memory/p2p_copy_coherency.cpp +++ b/tests/src/runtimeApi/memory/p2p_copy_coherency.cpp @@ -36,7 +36,6 @@ THE SOFTWARE. #define USE_HCC_MEMTRACKER 0 #endif -#define USE_HSA_COPY 1 int elementSizes[] = {16, 1024,524288}; int nSizes = sizeof(elementSizes) / sizeof(int); @@ -102,11 +101,8 @@ void runTest(bool stepAIsCopy, bool hostSync, hipStream_t gpu0Stream, hipStream_ hipStream_t stepAStream = gpu0Stream; if (stepAIsCopy) { -#ifdef USE_HSA_COPY HIPCHECK(hipMemcpyAsync(dataGpu1, dataGpu0_0, sizeElements, hipMemcpyDeviceToDevice, stepAStream)); -#endif } else { - //assert(0); // not yet supported. unsigned blocks = HipTest::setNumBlocks(blocksPerCU, threadsPerBlock, numElements); hipLaunchKernelGGL(memcpyIntKernel, dim3(blocks), dim3(threadsPerBlock), 0, gpu0Stream, dataGpu0_0, dataGpu1, numElements);