Clean up old work-week and USE_* refs
Change-Id: I929c979fa085f8e5205194cbccca46e9b5516aa9
This commit is contained in:
@@ -25,7 +25,7 @@ THE SOFTWARE.
|
||||
#include "hip/hcc_detail/staging_buffer.h"
|
||||
|
||||
|
||||
#if defined(__HCC__) && (__hcc_workweek__ < 16155)
|
||||
#if defined(__HCC__) && (__hcc_workweek__ < 16186)
|
||||
#error("This version of HIP requires a newer version of HCC.");
|
||||
#endif
|
||||
|
||||
@@ -37,11 +37,7 @@ THE SOFTWARE.
|
||||
// Compile peer-to-peer support.
|
||||
// >= 2 : use HCC hc:accelerator::get_is_peer
|
||||
// >= 3 : use hc::am_memtracker_update_peers(...)
|
||||
#define USE_PEER_TO_PEER 2
|
||||
|
||||
|
||||
// Use new am_memory_host_lock APIs:
|
||||
#define USE_HCC_LOCK_API 1
|
||||
#define USE_PEER_TO_PEER 3
|
||||
|
||||
|
||||
//---
|
||||
|
||||
@@ -134,11 +134,7 @@ hipError_t hipEventSynchronize(hipEvent_t event)
|
||||
device->locked_syncDefaultStream(true);
|
||||
return ihipLogStatus(hipSuccess);
|
||||
} else {
|
||||
#if __hcc_workweek__ >= 16033
|
||||
eh->_marker.wait((eh->_flags & hipEventBlockingSync) ? hc::hcWaitModeBlocked : hc::hcWaitModeActive);
|
||||
#else
|
||||
eh->_marker.wait();
|
||||
#endif
|
||||
eh->_stream->locked_reclaimSignals(eh->_copy_seq_id);
|
||||
|
||||
return ihipLogStatus(hipSuccess);
|
||||
|
||||
@@ -20,7 +20,6 @@ THE SOFTWARE.
|
||||
#include <hc.hpp>
|
||||
#include"hcc_detail/hip_ldg.h"
|
||||
|
||||
#if __hcc_workweek__ >= 16164
|
||||
__device__ char __ldg(const char* ptr)
|
||||
{
|
||||
return ptr[0];
|
||||
@@ -167,6 +166,5 @@ __device__ double2 __ldg(const double2* ptr)
|
||||
return ptr[0];
|
||||
}
|
||||
|
||||
#endif
|
||||
|
||||
|
||||
|
||||
@@ -254,11 +254,7 @@ hipError_t hipHostRegister(void *hostPtr, size_t sizeBytes, unsigned int flags)
|
||||
for(int i=0;i<g_deviceCnt;i++){
|
||||
vecAcc.push_back(g_devices[i]._acc);
|
||||
}
|
||||
#if USE_HCC_LOCK_API
|
||||
am_status = hc::am_memory_host_lock(device->_acc, hostPtr, sizeBytes, &vecAcc[0], vecAcc.size());
|
||||
#else
|
||||
am_status = AM_ERROR_MISC;
|
||||
#endif
|
||||
if(am_status == AM_SUCCESS){
|
||||
hip_status = hipSuccess;
|
||||
}else{
|
||||
@@ -281,11 +277,7 @@ hipError_t hipHostUnregister(void *hostPtr)
|
||||
if(hostPtr == NULL){
|
||||
hip_status = hipErrorInvalidValue;
|
||||
}else{
|
||||
#if USE_HCC_LOCK_API
|
||||
am_status_t am_status = hc::am_memory_host_unlock(device->_acc, hostPtr);
|
||||
#else
|
||||
am_status_t am_status = AM_ERROR_MISC;
|
||||
#endif
|
||||
if(am_status != AM_SUCCESS){
|
||||
hip_status = hipErrorHostMemoryNotRegistered;
|
||||
}
|
||||
|
||||
Reference in New Issue
Block a user