AI/rocm-systems
2
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

P4 to Git Change 1593706 by skudchad@skudchad_test2_win_opencl on 2018/08/14 18:44:29

Browse Source 
SWDEV-145570 - [HIP] Implement hipError*

	ReviewBoardURL = http://ocltc.amd.com/reviews/r/15619/diff/

Affected files ...

... //depot/stg/opencl/drivers/opencl/api/hip/hip_context.cpp#13 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_device.cpp#14 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_device_runtime.cpp#11 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_error.cpp#2 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_event.cpp#5 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_internal.hpp#13 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_memory.cpp#41 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_module.cpp#17 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_peer.cpp#2 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_platform.cpp#17 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_profile.cpp#2 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_stream.cpp#12 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_surface.cpp#2 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_texture.cpp#9 edit


[ROCm/hip commit: 21559762b8]
This commit is contained in:
foreman
2018-08-14 18:54:13 -04:00
parent 4a16441933
commit 3c4e2d4529
14 changed files with 396 additions and 265 deletions
Download Patch File Download Diff File Expand all files Collapse all files
projects/hip/api/hip/hip_context.cpp
+21 -20
View File
@@ -31,6 +31,7 @@ namespace hip {
thread_local amd::Context* g_context = nullptr; thread_local amd::Context* g_context = nullptr;
thread_local std::stack<amd::Context*> g_ctxtStack; thread_local std::stack<amd::Context*> g_ctxtStack;
thread_local hipError_t g_lastError = hipSuccess;
std::once_flag g_ihipInitialized; std::once_flag g_ihipInitialized;
std::map<amd::Context*, amd::HostQueue*> g_nullStreams; std::map<amd::Context*, amd::HostQueue*> g_nullStreams;
@@ -85,14 +86,14 @@ using namespace hip;
hipError_t hipInit(unsigned int flags) { hipError_t hipInit(unsigned int flags) {
HIP_INIT_API(flags); HIP_INIT_API(flags);
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipCtxCreate(hipCtx_t *ctx, unsigned int flags, hipDevice_t device) { hipError_t hipCtxCreate(hipCtx_t *ctx, unsigned int flags, hipDevice_t device) {
HIP_INIT_API(ctx, flags, device); HIP_INIT_API(ctx, flags, device);
if (static_cast<size_t>(device) >= g_devices.size()) { if (static_cast<size_t>(device) >= g_devices.size()) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
*ctx = reinterpret_cast<hipCtx_t>(g_devices[device]); *ctx = reinterpret_cast<hipCtx_t>(g_devices[device]);
@@ -101,7 +102,7 @@ hipError_t hipCtxCreate(hipCtx_t *ctx, unsigned int flags, hipDevice_t device)
g_devices[device]->retain(); g_devices[device]->retain();
g_ctxtStack.push(g_devices[device]); g_ctxtStack.push(g_devices[device]);
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipCtxSetCurrent(hipCtx_t ctx) { hipError_t hipCtxSetCurrent(hipCtx_t ctx) {
@@ -119,7 +120,7 @@ hipError_t hipCtxSetCurrent(hipCtx_t ctx) {
g_ctxtStack.push(hip::getCurrentContext()); g_ctxtStack.push(hip::getCurrentContext());
} }
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipCtxGetCurrent(hipCtx_t* ctx) { hipError_t hipCtxGetCurrent(hipCtx_t* ctx) {
@@ -127,19 +128,19 @@ hipError_t hipCtxGetCurrent(hipCtx_t* ctx) {
*ctx = reinterpret_cast<hipCtx_t>(hip::getCurrentContext()); *ctx = reinterpret_cast<hipCtx_t>(hip::getCurrentContext());
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipRuntimeGetVersion(int *runtimeVersion) { hipError_t hipRuntimeGetVersion(int *runtimeVersion) {
HIP_INIT_API(runtimeVersion); HIP_INIT_API(runtimeVersion);
if (!runtimeVersion) { if (!runtimeVersion) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
*runtimeVersion = AMD_PLATFORM_BUILD_NUMBER; *runtimeVersion = AMD_PLATFORM_BUILD_NUMBER;
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipCtxDestroy(hipCtx_t ctx) { hipError_t hipCtxDestroy(hipCtx_t ctx) {
@@ -147,7 +148,7 @@ hipError_t hipCtxDestroy(hipCtx_t ctx) {
amd::Context* amdContext = reinterpret_cast<amd::Context*>(as_amd(ctx)); amd::Context* amdContext = reinterpret_cast<amd::Context*>(as_amd(ctx));
if (amdContext == nullptr) { if (amdContext == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
// Need to remove the ctx of calling thread if its the top one // Need to remove the ctx of calling thread if its the top one
@@ -163,7 +164,7 @@ hipError_t hipCtxDestroy(hipCtx_t ctx) {
} }
} }
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipCtxPopCurrent(hipCtx_t* ctx) { hipError_t hipCtxPopCurrent(hipCtx_t* ctx) {
@@ -171,17 +172,17 @@ hipError_t hipCtxPopCurrent(hipCtx_t* ctx) {
amd::Context* amdContext = reinterpret_cast<amd::Context*>(as_amd(ctx)); amd::Context* amdContext = reinterpret_cast<amd::Context*>(as_amd(ctx));
if (amdContext == nullptr) { if (amdContext == nullptr) {
return hipErrorInvalidContext; HIP_RETURN(hipErrorInvalidContext);
} }
if (!g_ctxtStack.empty()) { if (!g_ctxtStack.empty()) {
amdContext = g_ctxtStack.top(); amdContext = g_ctxtStack.top();
g_ctxtStack.pop(); g_ctxtStack.pop();
} else { } else {
return hipErrorInvalidContext; HIP_RETURN(hipErrorInvalidContext);
} }
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipCtxPushCurrent(hipCtx_t ctx) { hipError_t hipCtxPushCurrent(hipCtx_t ctx) {
@@ -189,13 +190,13 @@ hipError_t hipCtxPushCurrent(hipCtx_t ctx) {
amd::Context* amdContext = reinterpret_cast<amd::Context*>(as_amd(ctx)); amd::Context* amdContext = reinterpret_cast<amd::Context*>(as_amd(ctx));
if (amdContext == nullptr) { if (amdContext == nullptr) {
return hipErrorInvalidContext; HIP_RETURN(hipErrorInvalidContext);
} }
hip::g_context = amdContext; hip::g_context = amdContext;
g_ctxtStack.push(hip::getCurrentContext()); g_ctxtStack.push(hip::getCurrentContext());
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipDriverGetVersion(int* driverVersion) { hipError_t hipDriverGetVersion(int* driverVersion) {
@@ -205,13 +206,13 @@ hipError_t hipDriverGetVersion(int* driverVersion) {
const auto& info = deviceHandle->info(); const auto& info = deviceHandle->info();
if (driverVersion) { if (driverVersion) {
*driverVersion = AMD_PLATFORM_BUILD_NUMBER * 100 + *driverVersion = AMD_PLATFORM_BUILD_NUMBER * 100 +
AMD_PLATFORM_REVISION_NUMBER; AMD_PLATFORM_REVISION_NUMBER;
} else { } else {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
return hipSuccess;; HIP_RETURN(hipSuccess);
} }
hipError_t hipCtxGetDevice(hipDevice_t* device) { hipError_t hipCtxGetDevice(hipDevice_t* device) {
@@ -221,11 +222,11 @@ hipError_t hipCtxGetDevice(hipDevice_t* device) {
for (unsigned int i = 0; i < g_devices.size(); i++) { for (unsigned int i = 0; i < g_devices.size(); i++) {
if (g_devices[i] == hip::getCurrentContext()) { if (g_devices[i] == hip::getCurrentContext()) {
*device = static_cast<hipDevice_t>(i); *device = static_cast<hipDevice_t>(i);
return hipSuccess; HIP_RETURN(hipSuccess);
} }
} }
} else { } else {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
return hipErrorUnknown; return hipErrorUnknown;
projects/hip/api/hip/hip_device.cpp
+18 -18
View File
@@ -30,10 +30,10 @@ hipError_t hipDeviceGet(hipDevice_t *device, int deviceId) {
if (device != nullptr) { if (device != nullptr) {
*device = deviceId; *device = deviceId;
} else { } else {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
return hipSuccess; HIP_RETURN(hipSuccess);
}; };
hipError_t hipFuncSetCacheConfig (const void* func, hipFuncCache_t cacheConfig) { hipError_t hipFuncSetCacheConfig (const void* func, hipFuncCache_t cacheConfig) {
@@ -42,7 +42,7 @@ hipError_t hipFuncSetCacheConfig (const void* func, hipFuncCache_t cacheConfig)
// No way to set cache config yet. // No way to set cache config yet.
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipDeviceTotalMem (size_t *bytes, hipDevice_t device) { hipError_t hipDeviceTotalMem (size_t *bytes, hipDevice_t device) {
@@ -50,11 +50,11 @@ hipError_t hipDeviceTotalMem (size_t *bytes, hipDevice_t device) {
HIP_INIT_API(bytes, device); HIP_INIT_API(bytes, device);
if (device < 0 || static_cast<size_t>(device) >= g_devices.size()) { if (device < 0 || static_cast<size_t>(device) >= g_devices.size()) {
return hipErrorInvalidDevice; HIP_RETURN(hipErrorInvalidDevice);
} }
if (bytes == nullptr) { if (bytes == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
auto* deviceHandle = g_devices[device]->devices()[0]; auto* deviceHandle = g_devices[device]->devices()[0];
@@ -62,7 +62,7 @@ hipError_t hipDeviceTotalMem (size_t *bytes, hipDevice_t device) {
*bytes = info.globalMemSize_; *bytes = info.globalMemSize_;
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipDeviceComputeCapability(int *major, int *minor, hipDevice_t device) { hipError_t hipDeviceComputeCapability(int *major, int *minor, hipDevice_t device) {
@@ -70,11 +70,11 @@ hipError_t hipDeviceComputeCapability(int *major, int *minor, hipDevice_t device
HIP_INIT_API(major, minor, device); HIP_INIT_API(major, minor, device);
if (device < 0 || static_cast<size_t>(device) >= g_devices.size()) { if (device < 0 || static_cast<size_t>(device) >= g_devices.size()) {
return hipErrorInvalidDevice; HIP_RETURN(hipErrorInvalidDevice);
} }
if (major == nullptr || minor == nullptr) { if (major == nullptr || minor == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
auto* deviceHandle = g_devices[device]->devices()[0]; auto* deviceHandle = g_devices[device]->devices()[0];
@@ -82,13 +82,13 @@ hipError_t hipDeviceComputeCapability(int *major, int *minor, hipDevice_t device
*major = info.gfxipVersion_ / 100; *major = info.gfxipVersion_ / 100;
*minor = info.gfxipVersion_ % 100; *minor = info.gfxipVersion_ % 100;
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipDeviceGetCount(int* count) { hipError_t hipDeviceGetCount(int* count) {
HIP_INIT_API(count); HIP_INIT_API(count);
return ihipDeviceGetCount(count); HIP_RETURN(ihipDeviceGetCount(count));
} }
hipError_t ihipDeviceGetCount(int* count) { hipError_t ihipDeviceGetCount(int* count) {
@@ -107,11 +107,11 @@ hipError_t hipDeviceGetName(char *name, int len, hipDevice_t device) {
HIP_INIT_API((void*)name, len, device); HIP_INIT_API((void*)name, len, device);
if (device < 0 || static_cast<size_t>(device) >= g_devices.size()) { if (device < 0 || static_cast<size_t>(device) >= g_devices.size()) {
return hipErrorInvalidDevice; HIP_RETURN(hipErrorInvalidDevice);
} }
if (name == nullptr) { if (name == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
auto* deviceHandle = g_devices[device]->devices()[0]; auto* deviceHandle = g_devices[device]->devices()[0];
@@ -120,18 +120,18 @@ hipError_t hipDeviceGetName(char *name, int len, hipDevice_t device) {
len = ((cl_uint)len < ::strlen(info.boardName_)) ? len : 128; len = ((cl_uint)len < ::strlen(info.boardName_)) ? len : 128;
::strncpy(name, info.boardName_, len); ::strncpy(name, info.boardName_, len);
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipGetDeviceProperties ( hipDeviceProp_t* props, hipDevice_t device ) { hipError_t hipGetDeviceProperties ( hipDeviceProp_t* props, hipDevice_t device ) {
HIP_INIT_API(props, device); HIP_INIT_API(props, device);
if (props == nullptr) { if (props == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
if (unsigned(device) >= g_devices.size()) { if (unsigned(device) >= g_devices.size()) {
return hipErrorInvalidDevice; HIP_RETURN(hipErrorInvalidDevice);
} }
auto* deviceHandle = g_devices[device]->devices()[0]; auto* deviceHandle = g_devices[device]->devices()[0];
@@ -188,7 +188,7 @@ hipError_t hipGetDeviceProperties ( hipDeviceProp_t* props, hipDevice_t device )
deviceProps.gcnArch = info.gfxipVersion_; deviceProps.gcnArch = info.gfxipVersion_;
*props = deviceProps; *props = deviceProps;
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipHccGetAccelerator(int deviceId, hc::accelerator* acc) { hipError_t hipHccGetAccelerator(int deviceId, hc::accelerator* acc) {
@@ -196,7 +196,7 @@ hipError_t hipHccGetAccelerator(int deviceId, hc::accelerator* acc) {
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipHccGetAcceleratorView(hipStream_t stream, hc::accelerator_view** av) { hipError_t hipHccGetAcceleratorView(hipStream_t stream, hc::accelerator_view** av) {
@@ -204,5 +204,5 @@ hipError_t hipHccGetAcceleratorView(hipStream_t stream, hc::accelerator_view** a
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
projects/hip/api/hip/hip_device_runtime.cpp
+38 -38
View File
@@ -29,7 +29,7 @@ hipError_t hipChooseDevice(int* device, const hipDeviceProp_t* properties) {
HIP_INIT_API(device, properties); HIP_INIT_API(device, properties);
if (device == nullptr || properties == nullptr) { if (device == nullptr || properties == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
*device = 0; *device = 0;
@@ -135,7 +135,7 @@ hipError_t hipChooseDevice(int* device, const hipDeviceProp_t* properties) {
} }
} }
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipDeviceGetAttribute(int* pi, hipDeviceAttribute_t attr, int device) { hipError_t hipDeviceGetAttribute(int* pi, hipDeviceAttribute_t attr, int device) {
@@ -143,19 +143,19 @@ hipError_t hipDeviceGetAttribute(int* pi, hipDeviceAttribute_t attr, int device)
HIP_INIT_API(pi, attr, device); HIP_INIT_API(pi, attr, device);
if (pi == nullptr) { if (pi == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
int count = 0; int count = 0;
ihipDeviceGetCount(&count); ihipDeviceGetCount(&count);
if (device < 0 || device >= count) { if (device < 0 || device >= count) {
return hipErrorInvalidDevice; HIP_RETURN(hipErrorInvalidDevice);
} }
//FIXME: should we cache the props, or just select from deviceHandle->info_? //FIXME: should we cache the props, or just select from deviceHandle->info_?
hipDeviceProp_t prop = {0}; hipDeviceProp_t prop = {0};
hipError_t err = hipGetDeviceProperties(&prop, device); hipError_t err = hipGetDeviceProperties(&prop, device);
if (err != hipSuccess) return err; if (err != hipSuccess) HIP_RETURN(err);
switch (attr) { switch (attr) {
case hipDeviceAttributeMaxThreadsPerBlock: case hipDeviceAttributeMaxThreadsPerBlock:
@@ -234,10 +234,10 @@ hipError_t hipDeviceGetAttribute(int* pi, hipDeviceAttribute_t attr, int device)
*pi = prop.isMultiGpuBoard; *pi = prop.isMultiGpuBoard;
break; break;
default: default:
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipDeviceGetByPCIBusId(int* device, const char*pciBusIdstr) { hipError_t hipDeviceGetByPCIBusId(int* device, const char*pciBusIdstr) {
@@ -245,7 +245,7 @@ hipError_t hipDeviceGetByPCIBusId(int* device, const char*pciBusIdstr) {
HIP_INIT_API(device, pciBusIdstr); HIP_INIT_API(device, pciBusIdstr);
if (device == nullptr || pciBusIdstr == nullptr) { if (device == nullptr || pciBusIdstr == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
int pciBusID = -1; int pciBusID = -1;
@@ -268,19 +268,19 @@ hipError_t hipDeviceGetByPCIBusId(int* device, const char*pciBusIdstr) {
} }
} }
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipDeviceGetCacheConfig ( hipFuncCache_t * cacheConfig ) { hipError_t hipDeviceGetCacheConfig ( hipFuncCache_t * cacheConfig ) {
HIP_INIT_API(cacheConfig); HIP_INIT_API(cacheConfig);
if(cacheConfig == nullptr) { if(cacheConfig == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
*cacheConfig = hipFuncCache_t(); *cacheConfig = hipFuncCache_t();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipDeviceGetLimit ( size_t* pValue, hipLimit_t limit ) { hipError_t hipDeviceGetLimit ( size_t* pValue, hipLimit_t limit ) {
@@ -288,23 +288,23 @@ hipError_t hipDeviceGetLimit ( size_t* pValue, hipLimit_t limit ) {
HIP_INIT_API(pValue, limit); HIP_INIT_API(pValue, limit);
if(pValue == nullptr) { if(pValue == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
if(limit == hipLimitMallocHeapSize) { if(limit == hipLimitMallocHeapSize) {
hipDeviceProp_t prop; hipDeviceProp_t prop;
hipGetDeviceProperties(&prop, 0); hipGetDeviceProperties(&prop, 0);
*pValue = prop.totalGlobalMem; *pValue = prop.totalGlobalMem;
return hipSuccess; HIP_RETURN(hipSuccess);
} else { } else {
return hipErrorUnsupportedLimit; HIP_RETURN(hipErrorUnsupportedLimit);
} }
} }
/** /**
hipError_t hipDeviceGetP2PAttribute ( int* value, hipDeviceP2PAttr attr, int srcDevice, int dstDevice ) { hipError_t hipDeviceGetP2PAttribute ( int* value, hipDeviceP2PAttr attr, int srcDevice, int dstDevice ) {
assert(0); assert(0);
return hipSuccess; HIP_RETURN(hipSuccess);
} }
**/ **/
@@ -315,11 +315,11 @@ hipError_t hipDeviceGetPCIBusId ( char* pciBusId, int len, int device ) {
int count; int count;
ihipDeviceGetCount(&count); ihipDeviceGetCount(&count);
if (device < 0 || device > count) { if (device < 0 || device > count) {
return hipErrorInvalidDevice; HIP_RETURN(hipErrorInvalidDevice);
} }
if (pciBusId == nullptr || len < 0) { if (pciBusId == nullptr || len < 0) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
hipDeviceProp_t prop; hipDeviceProp_t prop;
@@ -330,7 +330,7 @@ hipError_t hipDeviceGetPCIBusId ( char* pciBusId, int len, int device ) {
prop.pciBusID, prop.pciBusID,
prop.pciDeviceID); prop.pciDeviceID);
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipDeviceGetSharedMemConfig ( hipSharedMemConfig * pConfig ) { hipError_t hipDeviceGetSharedMemConfig ( hipSharedMemConfig * pConfig ) {
@@ -338,12 +338,12 @@ hipError_t hipDeviceGetSharedMemConfig ( hipSharedMemConfig * pConfig ) {
*pConfig = hipSharedMemBankSizeFourByte; *pConfig = hipSharedMemBankSizeFourByte;
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipDeviceGetStreamPriorityRange ( int* leastPriority, int* greatestPriority ) { hipError_t hipDeviceGetStreamPriorityRange ( int* leastPriority, int* greatestPriority ) {
assert(0); assert(0);
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipDeviceReset ( void ) { hipError_t hipDeviceReset ( void ) {
@@ -351,7 +351,7 @@ hipError_t hipDeviceReset ( void ) {
/* FIXME */ /* FIXME */
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipDeviceSetCacheConfig ( hipFuncCache_t cacheConfig ) { hipError_t hipDeviceSetCacheConfig ( hipFuncCache_t cacheConfig ) {
@@ -359,11 +359,11 @@ hipError_t hipDeviceSetCacheConfig ( hipFuncCache_t cacheConfig ) {
// No way to set cache config yet. // No way to set cache config yet.
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipDeviceSetLimit ( hipLimit_t limit, size_t value ) { hipError_t hipDeviceSetLimit ( hipLimit_t limit, size_t value ) {
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipDeviceSetSharedMemConfig ( hipSharedMemConfig config ) { hipError_t hipDeviceSetSharedMemConfig ( hipSharedMemConfig config ) {
@@ -371,7 +371,7 @@ hipError_t hipDeviceSetSharedMemConfig ( hipSharedMemConfig config ) {
// No way to set cache config yet. // No way to set cache config yet.
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipDeviceSynchronize ( void ) { hipError_t hipDeviceSynchronize ( void ) {
@@ -382,11 +382,11 @@ hipError_t hipDeviceSynchronize ( void ) {
amd::HostQueue* queue = hip::getNullStream(); amd::HostQueue* queue = hip::getNullStream();
if (!queue) { if (!queue) {
return hipErrorOutOfMemory; HIP_RETURN(hipErrorOutOfMemory);
} }
queue->finish(); queue->finish();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipGetDevice ( int* deviceId ) { hipError_t hipGetDevice ( int* deviceId ) {
@@ -396,24 +396,24 @@ hipError_t hipGetDevice ( int* deviceId ) {
for (unsigned int i = 0; i < g_devices.size(); i++) { for (unsigned int i = 0; i < g_devices.size(); i++) {
if (g_devices[i] == hip::getCurrentContext()) { if (g_devices[i] == hip::getCurrentContext()) {
*deviceId = i; *deviceId = i;
return hipSuccess; HIP_RETURN(hipSuccess);
} }
} }
} else { } else {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipGetDeviceCount ( int* count ) { hipError_t hipGetDeviceCount ( int* count ) {
HIP_INIT_API(count); HIP_INIT_API(count);
return ihipDeviceGetCount(count); HIP_RETURN(ihipDeviceGetCount(count));
} }
hipError_t hipGetDeviceFlags ( unsigned int* flags ) { hipError_t hipGetDeviceFlags ( unsigned int* flags ) {
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipIpcGetEventHandle ( hipIpcEventHandle_t* handle, hipEvent_t event ) { hipError_t hipIpcGetEventHandle ( hipIpcEventHandle_t* handle, hipEvent_t event ) {
@@ -421,7 +421,7 @@ hipError_t hipIpcGetEventHandle ( hipIpcEventHandle_t* handle, hipEvent_t event
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipIpcOpenEventHandle ( hipEvent_t* event, hipIpcEventHandle_t handle ) { hipError_t hipIpcOpenEventHandle ( hipEvent_t* event, hipIpcEventHandle_t handle ) {
@@ -429,7 +429,7 @@ hipError_t hipIpcOpenEventHandle ( hipEvent_t* event, hipIpcEventHandle_t handle
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipSetDevice ( int device ) { hipError_t hipSetDevice ( int device ) {
@@ -438,9 +438,9 @@ hipError_t hipSetDevice ( int device ) {
if (static_cast<unsigned int>(device) < g_devices.size()) { if (static_cast<unsigned int>(device) < g_devices.size()) {
hip::setCurrentContext(device); hip::setCurrentContext(device);
return hipSuccess; HIP_RETURN(hipSuccess);
} }
return hipErrorInvalidDevice; HIP_RETURN(hipErrorInvalidDevice);
} }
hipError_t hipSetDeviceFlags ( unsigned int flags ) { hipError_t hipSetDeviceFlags ( unsigned int flags ) {
@@ -453,10 +453,10 @@ hipError_t hipSetDeviceFlags ( unsigned int flags ) {
hipDeviceScheduleMask | hipDeviceMapHost | hipDeviceLmemResizeToMax; hipDeviceScheduleMask | hipDeviceMapHost | hipDeviceLmemResizeToMax;
if (flags & (~supportedFlags)) { if (flags & (~supportedFlags)) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipSetValidDevices ( int* device_arr, int len ) { hipError_t hipSetValidDevices ( int* device_arr, int len ) {
@@ -464,6 +464,6 @@ hipError_t hipSetValidDevices ( int* device_arr, int len ) {
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
projects/hip/api/hip/hip_error.cpp
+128 -4
View File
@@ -27,22 +27,146 @@ THE SOFTWARE.
hipError_t hipGetLastError() hipError_t hipGetLastError()
{ {
HIP_INIT_API(); HIP_INIT_API();
return hipErrorUnknown; hipError_t err = hip::g_lastError;
hip::g_lastError = hipSuccess;
return err;
} }
hipError_t hipPeekAtLastError() hipError_t hipPeekAtLastError()
{ {
HIP_INIT_API(); HIP_INIT_API();
return hipErrorUnknown; hipError_t err = hip::g_lastError;
HIP_RETURN(err);
} }
const char *hipGetErrorName(hipError_t hip_error) const char *hipGetErrorName(hipError_t hip_error)
{ {
return ""; switch (hip_error) {
case hipSuccess:
return "hipSuccess";
case hipErrorOutOfMemory:
return "hipErrorOutOfMemory";
case hipErrorNotInitialized:
return "hipErrorNotInitialized";
case hipErrorDeinitialized:
return "hipErrorDeinitialized";
case hipErrorProfilerDisabled:
return "hipErrorProfilerDisabled";
case hipErrorProfilerNotInitialized:
return "hipErrorProfilerNotInitialized";
case hipErrorProfilerAlreadyStarted:
return "hipErrorProfilerAlreadyStarted";
case hipErrorProfilerAlreadyStopped:
return "hipErrorProfilerAlreadyStopped";
case hipErrorInvalidImage:
return "hipErrorInvalidImage";
case hipErrorInvalidContext:
return "hipErrorInvalidContext";
case hipErrorContextAlreadyCurrent:
return "hipErrorContextAlreadyCurrent";
case hipErrorMapFailed:
return "hipErrorMapFailed";
case hipErrorUnmapFailed:
return "hipErrorUnmapFailed";
case hipErrorArrayIsMapped:
return "hipErrorArrayIsMapped";
case hipErrorAlreadyMapped:
return "hipErrorAlreadyMapped";
case hipErrorNoBinaryForGpu:
return "hipErrorNoBinaryForGpu";
case hipErrorAlreadyAcquired:
return "hipErrorAlreadyAcquired";
case hipErrorNotMapped:
return "hipErrorNotMapped";
case hipErrorNotMappedAsArray:
return "hipErrorNotMappedAsArray";
case hipErrorNotMappedAsPointer:
return "hipErrorNotMappedAsPointer";
case hipErrorECCNotCorrectable:
return "hipErrorECCNotCorrectable";
case hipErrorUnsupportedLimit:
return "hipErrorUnsupportedLimit";
case hipErrorContextAlreadyInUse:
return "hipErrorContextAlreadyInUse";
case hipErrorPeerAccessUnsupported:
return "hipErrorPeerAccessUnsupported";
case hipErrorInvalidKernelFile:
return "hipErrorInvalidKernelFile";
case hipErrorInvalidGraphicsContext:
return "hipErrorInvalidGraphicsContext";
case hipErrorInvalidSource:
return "hipErrorInvalidSource";
case hipErrorFileNotFound:
return "hipErrorFileNotFound";
case hipErrorSharedObjectSymbolNotFound:
return "hipErrorSharedObjectSymbolNotFound";
case hipErrorSharedObjectInitFailed:
return "hipErrorSharedObjectInitFailed";
case hipErrorOperatingSystem:
return "hipErrorOperatingSystem";
case hipErrorSetOnActiveProcess:
return "hipErrorSetOnActiveProcess";
case hipErrorInvalidHandle:
return "hipErrorInvalidHandle";
case hipErrorNotFound:
return "hipErrorNotFound";
case hipErrorIllegalAddress:
return "hipErrorIllegalAddress";
case hipErrorMissingConfiguration:
return "hipErrorMissingConfiguration";
case hipErrorMemoryAllocation:
return "hipErrorMemoryAllocation";
case hipErrorInitializationError:
return "hipErrorInitializationError";
case hipErrorLaunchFailure:
return "hipErrorLaunchFailure";
case hipErrorPriorLaunchFailure:
return "hipErrorPriorLaunchFailure";
case hipErrorLaunchTimeOut:
return "hipErrorLaunchTimeOut";
case hipErrorLaunchOutOfResources:
return "hipErrorLaunchOutOfResources";
case hipErrorInvalidDeviceFunction:
return "hipErrorInvalidDeviceFunction";
case hipErrorInvalidConfiguration:
return "hipErrorInvalidConfiguration";
case hipErrorInvalidDevice:
return "hipErrorInvalidDevice";
case hipErrorInvalidValue:
return "hipErrorInvalidValue";
case hipErrorInvalidDevicePointer:
return "hipErrorInvalidDevicePointer";
case hipErrorInvalidMemcpyDirection:
return "hipErrorInvalidMemcpyDirection";
case hipErrorUnknown:
return "hipErrorUnknown";
case hipErrorInvalidResourceHandle:
return "hipErrorInvalidResourceHandle";
case hipErrorNotReady:
return "hipErrorNotReady";
case hipErrorNoDevice:
return "hipErrorNoDevice";
case hipErrorPeerAccessAlreadyEnabled:
return "hipErrorPeerAccessAlreadyEnabled";
case hipErrorPeerAccessNotEnabled:
return "hipErrorPeerAccessNotEnabled";
case hipErrorRuntimeMemory:
return "hipErrorRuntimeMemory";
case hipErrorRuntimeOther:
return "hipErrorRuntimeOther";
case hipErrorHostMemoryAlreadyRegistered:
return "hipErrorHostMemoryAlreadyRegistered";
case hipErrorHostMemoryNotRegistered:
return "hipErrorHostMemoryNotRegistered";
case hipErrorTbd:
return "hipErrorTbd";
default:
return "hipErrorUnknown";
};
} }
const char *hipGetErrorString(hipError_t hip_error) const char *hipGetErrorString(hipError_t hip_error)
{ {
return ""; return hipGetErrorName(hip_error);
} }
projects/hip/api/hip/hip_event.cpp
+17 -17
View File
@@ -64,38 +64,38 @@ hipError_t ihipEventQuery(hipEvent_t event) {
e->event_->notifyCmdQueue(); e->event_->notifyCmdQueue();
return (e->event_->status() == CL_COMPLETE)? hipSuccess : hipErrorNotReady; return (e->event_->status() == CL_COMPLETE) ? hipSuccess : hipErrorNotReady;
} }
hipError_t hipEventCreateWithFlags(hipEvent_t* event, unsigned flags) { hipError_t hipEventCreateWithFlags(hipEvent_t* event, unsigned flags) {
HIP_INIT_API(event, flags); HIP_INIT_API(event, flags);
return ihipEventCreateWithFlags(event, flags); HIP_RETURN(ihipEventCreateWithFlags(event, flags));
} }
hipError_t hipEventCreate(hipEvent_t* event) { hipError_t hipEventCreate(hipEvent_t* event) {
HIP_INIT_API(event); HIP_INIT_API(event);
return ihipEventCreateWithFlags(event, 0); HIP_RETURN(ihipEventCreateWithFlags(event, 0));
} }
hipError_t hipEventDestroy(hipEvent_t event) { hipError_t hipEventDestroy(hipEvent_t event) {
HIP_INIT_API(event); HIP_INIT_API(event);
if (event == nullptr) { if (event == nullptr) {
return hipErrorInvalidResourceHandle; HIP_RETURN(hipErrorInvalidResourceHandle);
} }
delete reinterpret_cast<hip::Event*>(event); delete reinterpret_cast<hip::Event*>(event);
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipEventElapsedTime(float *ms, hipEvent_t start, hipEvent_t stop) { hipError_t hipEventElapsedTime(float *ms, hipEvent_t start, hipEvent_t stop) {
HIP_INIT_API(ms, start, stop); HIP_INIT_API(ms, start, stop);
if (start == nullptr || stop == nullptr) { if (start == nullptr || stop == nullptr) {
return hipErrorInvalidResourceHandle; HIP_RETURN(hipErrorInvalidResourceHandle);
} }
hip::Event* eStart = reinterpret_cast<hip::Event*>(start); hip::Event* eStart = reinterpret_cast<hip::Event*>(start);
@@ -103,32 +103,32 @@ hipError_t hipEventElapsedTime(float *ms, hipEvent_t start, hipEvent_t stop) {
if (eStart->event_ == nullptr || if (eStart->event_ == nullptr ||
eStop->event_ == nullptr) { eStop->event_ == nullptr) {
return hipErrorInvalidResourceHandle; HIP_RETURN(hipErrorInvalidResourceHandle);
} }
if ((eStart->flags | eStop->flags) & hipEventDisableTiming) { if ((eStart->flags | eStop->flags) & hipEventDisableTiming) {
return hipErrorInvalidResourceHandle; HIP_RETURN(hipErrorInvalidResourceHandle);
} }
if (ihipEventQuery(start) == hipErrorNotReady || if (ihipEventQuery(start) == hipErrorNotReady ||
ihipEventQuery(stop) == hipErrorNotReady) { ihipEventQuery(stop) == hipErrorNotReady) {
return hipErrorNotReady; HIP_RETURN(hipErrorNotReady);
} }
if (ms == nullptr) { if (ms == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
*ms = static_cast<float>(eStop->event_->profilingInfo().submitted_ - eStart->event_->profilingInfo().submitted_)/1000000.f; *ms = static_cast<float>(eStop->event_->profilingInfo().submitted_ - eStart->event_->profilingInfo().submitted_)/1000000.f;
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipEventRecord(hipEvent_t event, hipStream_t stream) { hipError_t hipEventRecord(hipEvent_t event, hipStream_t stream) {
HIP_INIT_API(event, stream); HIP_INIT_API(event, stream);
if (event == nullptr) { if (event == nullptr) {
return hipErrorInvalidResourceHandle; HIP_RETURN(hipErrorInvalidResourceHandle);
} }
hip::Event* e = reinterpret_cast<hip::Event*>(event); hip::Event* e = reinterpret_cast<hip::Event*>(event);
@@ -148,29 +148,29 @@ hipError_t hipEventRecord(hipEvent_t event, hipStream_t stream) {
e->event_ = &command->event(); e->event_ = &command->event();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipEventSynchronize(hipEvent_t event) { hipError_t hipEventSynchronize(hipEvent_t event) {
HIP_INIT_API(event); HIP_INIT_API(event);
if (event == nullptr) { if (event == nullptr) {
return hipErrorInvalidResourceHandle; HIP_RETURN(hipErrorInvalidResourceHandle);
} }
hip::Event* e = reinterpret_cast<hip::Event*>(event); hip::Event* e = reinterpret_cast<hip::Event*>(event);
if (e->event_ == nullptr) { if (e->event_ == nullptr) {
return hipErrorInvalidResourceHandle; HIP_RETURN(hipErrorInvalidResourceHandle);
} }
e->event_->awaitCompletion(); e->event_->awaitCompletion();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipEventQuery(hipEvent_t event) { hipError_t hipEventQuery(hipEvent_t event) {
HIP_INIT_API(event); HIP_INIT_API(event);
return ihipEventQuery(event); HIP_RETURN(ihipEventQuery(event));
} }
projects/hip/api/hip/hip_internal.hpp
+6 -1
View File
@@ -39,7 +39,7 @@ THE SOFTWARE.
#define HIP_INIT_API(...) \ #define HIP_INIT_API(...) \
amd::Thread* thread = amd::Thread::current(); \ amd::Thread* thread = amd::Thread::current(); \
if (!CL_CHECK_THREAD(thread)) { \ if (!CL_CHECK_THREAD(thread)) { \
return hipErrorOutOfMemory; \ HIP_RETURN(hipErrorOutOfMemory); \
} \ } \
HIP_INIT(); HIP_INIT();
@@ -51,6 +51,7 @@ class accelerator_view;
namespace hip { namespace hip {
extern std::once_flag g_ihipInitialized; extern std::once_flag g_ihipInitialized;
extern thread_local amd::Context* g_context; extern thread_local amd::Context* g_context;
extern thread_local hipError_t g_lastError;
extern void init(); extern void init();
@@ -64,5 +65,9 @@ extern std::vector<amd::Context*> g_devices;
extern hipError_t ihipDeviceGetCount(int* count); extern hipError_t ihipDeviceGetCount(int* count);
extern amd::Memory* getMemoryObject(const void* ptr, size_t& offset); extern amd::Memory* getMemoryObject(const void* ptr, size_t& offset);
#define HIP_RETURN(ret) \
hip::g_lastError = ret; \
return ret; \
#endif // HIP_SRC_HIP_INTERNAL_H #endif // HIP_SRC_HIP_INTERNAL_H
projects/hip/api/hip/hip_memory.cpp
+86 -86
View File
@@ -151,14 +151,14 @@ hipError_t ihipMemset(void* dst, int value, size_t sizeBytes, amd::HostQueue& qu
hipError_t hipMalloc(void** ptr, size_t sizeBytes) { hipError_t hipMalloc(void** ptr, size_t sizeBytes) {
HIP_INIT_API(ptr, sizeBytes); HIP_INIT_API(ptr, sizeBytes);
return ihipMalloc(ptr, sizeBytes, 0); HIP_RETURN(ihipMalloc(ptr, sizeBytes, 0));
} }
hipError_t hipHostMalloc(void** ptr, size_t sizeBytes, unsigned int flags) { hipError_t hipHostMalloc(void** ptr, size_t sizeBytes, unsigned int flags) {
HIP_INIT_API(ptr, sizeBytes, flags); HIP_INIT_API(ptr, sizeBytes, flags);
if (ptr == nullptr) { if (ptr == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
*ptr = nullptr; *ptr = nullptr;
@@ -166,10 +166,10 @@ hipError_t hipHostMalloc(void** ptr, size_t sizeBytes, unsigned int flags) {
// can't have both Coherent and NonCoherent flags set at the same time // can't have both Coherent and NonCoherent flags set at the same time
if ((flags & coherentFlags) == coherentFlags) { if ((flags & coherentFlags) == coherentFlags) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
return ihipMalloc(ptr, sizeBytes, CL_MEM_SVM_FINE_GRAIN_BUFFER | (flags << 16)); HIP_RETURN(ihipMalloc(ptr, sizeBytes, CL_MEM_SVM_FINE_GRAIN_BUFFER | (flags << 16)));
} }
hipError_t hipFree(void* ptr) { hipError_t hipFree(void* ptr) {
@@ -177,9 +177,9 @@ hipError_t hipFree(void* ptr) {
hip::syncStreams(); hip::syncStreams();
hip::getNullStream()->finish(); hip::getNullStream()->finish();
amd::SvmBuffer::free(*hip::getCurrentContext(), ptr); amd::SvmBuffer::free(*hip::getCurrentContext(), ptr);
return hipSuccess; HIP_RETURN(hipSuccess);
} }
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
hipError_t hipMemcpy(void* dst, const void* src, size_t sizeBytes, hipMemcpyKind kind) { hipError_t hipMemcpy(void* dst, const void* src, size_t sizeBytes, hipMemcpyKind kind) {
@@ -187,7 +187,7 @@ hipError_t hipMemcpy(void* dst, const void* src, size_t sizeBytes, hipMemcpyKind
hip::syncStreams(); hip::syncStreams();
amd::HostQueue* queue = hip::getNullStream(); amd::HostQueue* queue = hip::getNullStream();
return ihipMemcpy(dst, src, sizeBytes, kind, *queue); HIP_RETURN(ihipMemcpy(dst, src, sizeBytes, kind, *queue));
} }
hipError_t hipMemsetAsync(void* dst, int value, size_t sizeBytes, hipStream_t stream) { hipError_t hipMemsetAsync(void* dst, int value, size_t sizeBytes, hipStream_t stream) {
@@ -203,7 +203,7 @@ hipError_t hipMemsetAsync(void* dst, int value, size_t sizeBytes, hipStream_t st
queue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue(); queue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue();
} }
return ihipMemset(dst, value, sizeBytes, *queue, true); HIP_RETURN(ihipMemset(dst, value, sizeBytes, *queue, true));
} }
hipError_t hipMemset(void* dst, int value, size_t sizeBytes) { hipError_t hipMemset(void* dst, int value, size_t sizeBytes) {
@@ -212,7 +212,7 @@ hipError_t hipMemset(void* dst, int value, size_t sizeBytes) {
hip::syncStreams(); hip::syncStreams();
amd::HostQueue* queue = hip::getNullStream(); amd::HostQueue* queue = hip::getNullStream();
return ihipMemset(dst, value, sizeBytes, *queue); HIP_RETURN(ihipMemset(dst, value, sizeBytes, *queue));
} }
hipError_t hipMemPtrGetInfo(void *ptr, size_t *size) { hipError_t hipMemPtrGetInfo(void *ptr, size_t *size) {
@@ -222,12 +222,12 @@ hipError_t hipMemPtrGetInfo(void *ptr, size_t *size) {
amd::Memory* svmMem = getMemoryObject(ptr, offset); amd::Memory* svmMem = getMemoryObject(ptr, offset);
if (svmMem == nullptr) { if (svmMem == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
*size = svmMem->getSize(); *size = svmMem->getSize();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipHostFree(void* ptr) { hipError_t hipHostFree(void* ptr) {
@@ -235,9 +235,9 @@ hipError_t hipHostFree(void* ptr) {
if (amd::SvmBuffer::malloced(ptr)) { if (amd::SvmBuffer::malloced(ptr)) {
amd::SvmBuffer::free(*hip::getCurrentContext(), ptr); amd::SvmBuffer::free(*hip::getCurrentContext(), ptr);
return hipSuccess; HIP_RETURN(hipSuccess);
} }
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
hipError_t hipFreeArray(hipArray* array) { hipError_t hipFreeArray(hipArray* array) {
@@ -245,9 +245,9 @@ hipError_t hipFreeArray(hipArray* array) {
if (amd::SvmBuffer::malloced(array->data)) { if (amd::SvmBuffer::malloced(array->data)) {
amd::SvmBuffer::free(*hip::getCurrentContext(), array->data); amd::SvmBuffer::free(*hip::getCurrentContext(), array->data);
return hipSuccess; HIP_RETURN(hipSuccess);
} }
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
hipError_t hipMemGetAddressRange(hipDeviceptr_t* pbase, size_t* psize, hipDeviceptr_t dptr) { hipError_t hipMemGetAddressRange(hipDeviceptr_t* pbase, size_t* psize, hipDeviceptr_t dptr) {
@@ -259,13 +259,13 @@ hipError_t hipMemGetAddressRange(hipDeviceptr_t* pbase, size_t* psize, hipDevice
amd::Memory* svmMem = getMemoryObject(ptr, offset); amd::Memory* svmMem = getMemoryObject(ptr, offset);
if (svmMem == nullptr) { if (svmMem == nullptr) {
return hipErrorInvalidDevicePointer; HIP_RETURN(hipErrorInvalidDevicePointer);
} }
*pbase = svmMem->getSvmPtr(); *pbase = svmMem->getSvmPtr();
*psize = svmMem->getSize(); *psize = svmMem->getSize();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipMemGetInfo(size_t* free, size_t* total) { hipError_t hipMemGetInfo(size_t* free, size_t* total) {
@@ -274,17 +274,17 @@ hipError_t hipMemGetInfo(size_t* free, size_t* total) {
size_t freeMemory[2]; size_t freeMemory[2];
amd::Device* device = hip::getCurrentContext()->devices()[0]; amd::Device* device = hip::getCurrentContext()->devices()[0];
if(device == nullptr) { if(device == nullptr) {
return hipErrorInvalidDevice; HIP_RETURN(hipErrorInvalidDevice);
} }
if(!device->globalFreeMemory(freeMemory)) { if(!device->globalFreeMemory(freeMemory)) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
*free = freeMemory[0] * Ki; *free = freeMemory[0] * Ki;
*total = device->info().globalMemSize_; *total = device->info().globalMemSize_;
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t ihipMallocPitch(void** ptr, size_t* pitch, size_t width, size_t height, size_t depth, hipError_t ihipMallocPitch(void** ptr, size_t* pitch, size_t width, size_t height, size_t depth,
@@ -325,7 +325,7 @@ hipError_t hipMallocPitch(void** ptr, size_t* pitch, size_t width, size_t height
HIP_INIT_API(ptr, pitch, width, height); HIP_INIT_API(ptr, pitch, width, height);
const cl_image_format image_format = { CL_R, CL_UNSIGNED_INT8 }; const cl_image_format image_format = { CL_R, CL_UNSIGNED_INT8 };
return ihipMallocPitch(ptr, pitch, width, height, 1, CL_MEM_OBJECT_IMAGE2D, &image_format); HIP_RETURN(ihipMallocPitch(ptr, pitch, width, height, 1, CL_MEM_OBJECT_IMAGE2D, &image_format));
} }
hipError_t hipMalloc3D(hipPitchedPtr* pitchedDevPtr, hipExtent extent) { hipError_t hipMalloc3D(hipPitchedPtr* pitchedDevPtr, hipExtent extent) {
@@ -334,7 +334,7 @@ hipError_t hipMalloc3D(hipPitchedPtr* pitchedDevPtr, hipExtent extent) {
size_t pitch = 0; size_t pitch = 0;
if (pitchedDevPtr == nullptr) { if (pitchedDevPtr == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
const cl_image_format image_format = { CL_R, CL_UNSIGNED_INT8 }; const cl_image_format image_format = { CL_R, CL_UNSIGNED_INT8 };
@@ -348,7 +348,7 @@ hipError_t hipMalloc3D(hipPitchedPtr* pitchedDevPtr, hipExtent extent) {
pitchedDevPtr->ysize = extent.height; pitchedDevPtr->ysize = extent.height;
} }
return status; HIP_RETURN(status);
} }
hipError_t hipMemset3D(hipPitchedPtr pitchedDevPtr, int value, hipExtent extent) { hipError_t hipMemset3D(hipPitchedPtr pitchedDevPtr, int value, hipExtent extent) {
@@ -360,14 +360,14 @@ hipError_t hipMemset3D(hipPitchedPtr pitchedDevPtr, int value, hipExtent extent)
hip::syncStreams(); hip::syncStreams();
amd::HostQueue* queue = hip::getNullStream(); amd::HostQueue* queue = hip::getNullStream();
return ihipMemset(dst, value, sizeBytes, *queue); HIP_RETURN(ihipMemset(dst, value, sizeBytes, *queue));
} }
hipError_t hipArrayCreate(hipArray** array, const HIP_ARRAY_DESCRIPTOR* pAllocateArray) { hipError_t hipArrayCreate(hipArray** array, const HIP_ARRAY_DESCRIPTOR* pAllocateArray) {
HIP_INIT_API(array, pAllocateArray); HIP_INIT_API(array, pAllocateArray);
if (array[0]->width == 0) { if (array[0]->width == 0) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
*array = (hipArray*)malloc(sizeof(hipArray)); *array = (hipArray*)malloc(sizeof(hipArray));
@@ -393,7 +393,7 @@ hipError_t hipArrayCreate(hipArray** array, const HIP_ARRAY_DESCRIPTOR* pAllocat
hipError_t status = ihipMallocPitch(ptr, &pitch, array[0]->width, array[0]->height, 1, CL_MEM_OBJECT_IMAGE2D, hipError_t status = ihipMallocPitch(ptr, &pitch, array[0]->width, array[0]->height, 1, CL_MEM_OBJECT_IMAGE2D,
&image_format); &image_format);
return status; HIP_RETURN(status);
} }
hipError_t hipMallocArray(hipArray** array, const hipChannelFormatDesc* desc, hipError_t hipMallocArray(hipArray** array, const hipChannelFormatDesc* desc,
@@ -401,7 +401,7 @@ hipError_t hipMallocArray(hipArray** array, const hipChannelFormatDesc* desc,
HIP_INIT_API(array, desc, width, height, flags); HIP_INIT_API(array, desc, width, height, flags);
if (width == 0) { if (width == 0) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
*array = (hipArray*)malloc(sizeof(hipArray)); *array = (hipArray*)malloc(sizeof(hipArray));
@@ -436,7 +436,7 @@ hipError_t hipMallocArray(hipArray** array, const hipChannelFormatDesc* desc,
hipError_t status = ihipMallocPitch(ptr, &pitch, width, height, 1, CL_MEM_OBJECT_IMAGE2D, hipError_t status = ihipMallocPitch(ptr, &pitch, width, height, 1, CL_MEM_OBJECT_IMAGE2D,
&image_format); &image_format);
return status; HIP_RETURN(status);
} }
hipError_t hipMalloc3DArray(hipArray_t* array, const struct hipChannelFormatDesc* desc, hipError_t hipMalloc3DArray(hipArray_t* array, const struct hipChannelFormatDesc* desc,
@@ -475,7 +475,7 @@ hipError_t hipMalloc3DArray(hipArray_t* array, const struct hipChannelFormatDesc
hipError_t status = ihipMallocPitch(ptr, &pitch, extent.width, extent.height, extent.depth, hipError_t status = ihipMallocPitch(ptr, &pitch, extent.width, extent.height, extent.depth,
CL_MEM_OBJECT_IMAGE3D, &image_format); CL_MEM_OBJECT_IMAGE3D, &image_format);
return status; HIP_RETURN(status);
} }
hipError_t hipHostGetFlags(unsigned int* flagsPtr, void* hostPtr) { hipError_t hipHostGetFlags(unsigned int* flagsPtr, void* hostPtr) {
@@ -483,19 +483,19 @@ hipError_t hipHostGetFlags(unsigned int* flagsPtr, void* hostPtr) {
if (flagsPtr == nullptr || if (flagsPtr == nullptr ||
hostPtr == nullptr) { hostPtr == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
size_t offset = 0; size_t offset = 0;
amd::Memory* svmMem = getMemoryObject(hostPtr, offset); amd::Memory* svmMem = getMemoryObject(hostPtr, offset);
if (svmMem == nullptr) { if (svmMem == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
*flagsPtr = svmMem->getMemFlags() >> 16; *flagsPtr = svmMem->getMemFlags() >> 16;
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipHostRegister(void* hostPtr, size_t sizeBytes, unsigned int flags) { hipError_t hipHostRegister(void* hostPtr, size_t sizeBytes, unsigned int flags) {
@@ -506,12 +506,12 @@ hipError_t hipHostRegister(void* hostPtr, size_t sizeBytes, unsigned int flags)
if (!mem->create(hostPtr)) { if (!mem->create(hostPtr)) {
mem->release(); mem->release();
return hipErrorMemoryAllocation; HIP_RETURN(hipErrorMemoryAllocation);
} }
amd::MemObjMap::AddMemObj(hostPtr, mem); amd::MemObjMap::AddMemObj(hostPtr, mem);
return hipSuccess; HIP_RETURN(hipSuccess);
} else { } else {
return ihipMalloc(&hostPtr, sizeBytes, flags); HIP_RETURN(ihipMalloc(&hostPtr, sizeBytes, flags));
} }
} }
@@ -522,7 +522,7 @@ hipError_t hipHostUnregister(void* hostPtr) {
hip::syncStreams(); hip::syncStreams();
hip::getNullStream()->finish(); hip::getNullStream()->finish();
amd::SvmBuffer::free(*hip::getCurrentContext(), hostPtr); amd::SvmBuffer::free(*hip::getCurrentContext(), hostPtr);
return hipSuccess; HIP_RETURN(hipSuccess);
} else { } else {
size_t offset = 0; size_t offset = 0;
amd::Memory* mem = getMemoryObject(hostPtr, offset); amd::Memory* mem = getMemoryObject(hostPtr, offset);
@@ -532,16 +532,16 @@ hipError_t hipHostUnregister(void* hostPtr) {
hip::getNullStream()->finish(); hip::getNullStream()->finish();
amd::MemObjMap::RemoveMemObj(hostPtr); amd::MemObjMap::RemoveMemObj(hostPtr);
mem->release(); mem->release();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
} }
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
// Deprecated function: // Deprecated function:
hipError_t hipHostAlloc(void** ptr, size_t sizeBytes, unsigned int flags) { hipError_t hipHostAlloc(void** ptr, size_t sizeBytes, unsigned int flags) {
return ihipMalloc(ptr, sizeBytes, flags); HIP_RETURN(ihipMalloc(ptr, sizeBytes, flags));
}; };
@@ -551,7 +551,7 @@ hipError_t hipMemcpyToSymbol(const void* symbolName, const void* src, size_t cou
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipMemcpyFromSymbol(void* dst, const void* symbolName, size_t count, hipError_t hipMemcpyFromSymbol(void* dst, const void* symbolName, size_t count,
@@ -560,7 +560,7 @@ hipError_t hipMemcpyFromSymbol(void* dst, const void* symbolName, size_t count,
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipMemcpyToSymbolAsync(const void* symbolName, const void* src, size_t count, hipError_t hipMemcpyToSymbolAsync(const void* symbolName, const void* src, size_t count,
@@ -569,7 +569,7 @@ hipError_t hipMemcpyToSymbolAsync(const void* symbolName, const void* src, size_
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipMemcpyFromSymbolAsync(void* dst, const void* symbolName, size_t count, hipError_t hipMemcpyFromSymbolAsync(void* dst, const void* symbolName, size_t count,
@@ -578,7 +578,7 @@ hipError_t hipMemcpyFromSymbolAsync(void* dst, const void* symbolName, size_t co
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipMemcpyHtoD(hipDeviceptr_t dst, void* src, size_t sizeBytes) { hipError_t hipMemcpyHtoD(hipDeviceptr_t dst, void* src, size_t sizeBytes) {
@@ -587,7 +587,7 @@ hipError_t hipMemcpyHtoD(hipDeviceptr_t dst, void* src, size_t sizeBytes) {
hip::syncStreams(); hip::syncStreams();
amd::HostQueue* queue = hip::getNullStream(); amd::HostQueue* queue = hip::getNullStream();
return ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyHostToDevice, *queue); HIP_RETURN(ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyHostToDevice, *queue));
} }
hipError_t hipMemcpyDtoH(void* dst, hipDeviceptr_t src, size_t sizeBytes) { hipError_t hipMemcpyDtoH(void* dst, hipDeviceptr_t src, size_t sizeBytes) {
@@ -596,7 +596,7 @@ hipError_t hipMemcpyDtoH(void* dst, hipDeviceptr_t src, size_t sizeBytes) {
hip::syncStreams(); hip::syncStreams();
amd::HostQueue* queue = hip::getNullStream(); amd::HostQueue* queue = hip::getNullStream();
return ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyDeviceToHost, *queue); HIP_RETURN(ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyDeviceToHost, *queue));
} }
hipError_t hipMemcpyDtoD(hipDeviceptr_t dst, hipDeviceptr_t src, size_t sizeBytes) { hipError_t hipMemcpyDtoD(hipDeviceptr_t dst, hipDeviceptr_t src, size_t sizeBytes) {
@@ -605,7 +605,7 @@ hipError_t hipMemcpyDtoD(hipDeviceptr_t dst, hipDeviceptr_t src, size_t sizeByte
hip::syncStreams(); hip::syncStreams();
amd::HostQueue* queue = hip::getNullStream(); amd::HostQueue* queue = hip::getNullStream();
return ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyDeviceToDevice, *queue); HIP_RETURN(ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyDeviceToDevice, *queue));
} }
hipError_t hipMemcpyHtoH(void* dst, void* src, size_t sizeBytes) { hipError_t hipMemcpyHtoH(void* dst, void* src, size_t sizeBytes) {
@@ -614,7 +614,7 @@ hipError_t hipMemcpyHtoH(void* dst, void* src, size_t sizeBytes) {
hip::syncStreams(); hip::syncStreams();
amd::HostQueue* queue = hip::getNullStream(); amd::HostQueue* queue = hip::getNullStream();
return ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyHostToHost, *queue); HIP_RETURN(ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyHostToHost, *queue));
} }
hipError_t hipMemcpyAsync(void* dst, const void* src, size_t sizeBytes, hipError_t hipMemcpyAsync(void* dst, const void* src, size_t sizeBytes,
@@ -631,7 +631,7 @@ hipError_t hipMemcpyAsync(void* dst, const void* src, size_t sizeBytes,
queue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue(); queue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue();
} }
return ihipMemcpy(dst, src, sizeBytes, kind, *queue, true); HIP_RETURN(ihipMemcpy(dst, src, sizeBytes, kind, *queue, true));
} }
@@ -649,8 +649,8 @@ hipError_t hipMemcpyHtoDAsync(hipDeviceptr_t dst, void* src, size_t sizeBytes,
queue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue(); queue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue();
} }
return ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyHostToDevice, HIP_RETURN(ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyHostToDevice,
*queue, true); *queue, true));
} }
hipError_t hipMemcpyDtoDAsync(hipDeviceptr_t dst, hipDeviceptr_t src, size_t sizeBytes, hipError_t hipMemcpyDtoDAsync(hipDeviceptr_t dst, hipDeviceptr_t src, size_t sizeBytes,
@@ -667,8 +667,8 @@ hipError_t hipMemcpyDtoDAsync(hipDeviceptr_t dst, hipDeviceptr_t src, size_t siz
queue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue(); queue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue();
} }
return ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyDeviceToDevice, HIP_RETURN(ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyDeviceToDevice,
*queue, true); *queue, true));
} }
hipError_t hipMemcpyDtoHAsync(void* dst, hipDeviceptr_t src, size_t sizeBytes, hipError_t hipMemcpyDtoHAsync(void* dst, hipDeviceptr_t src, size_t sizeBytes,
@@ -685,8 +685,8 @@ hipError_t hipMemcpyDtoHAsync(void* dst, hipDeviceptr_t src, size_t sizeBytes,
queue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue(); queue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue();
} }
return ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyDeviceToHost, HIP_RETURN(ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyDeviceToHost,
*queue, true); *queue, true));
} }
hipError_t hipMemcpyParam2D(const hip_Memcpy2D* pCopy) { hipError_t hipMemcpyParam2D(const hip_Memcpy2D* pCopy) {
@@ -694,7 +694,7 @@ hipError_t hipMemcpyParam2D(const hip_Memcpy2D* pCopy) {
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t ihipMemcpy2D(void* dst, size_t dpitch, const void* src, size_t spitch, size_t width, hipError_t ihipMemcpy2D(void* dst, size_t dpitch, const void* src, size_t spitch, size_t width,
@@ -763,7 +763,7 @@ hipError_t hipMemcpy2D(void* dst, size_t dpitch, const void* src, size_t spitch,
hip::syncStreams(); hip::syncStreams();
amd::HostQueue* queue = hip::getNullStream(); amd::HostQueue* queue = hip::getNullStream();
return ihipMemcpy2D(dst, dpitch, src, spitch, width, height, kind, *queue); HIP_RETURN(ihipMemcpy2D(dst, dpitch, src, spitch, width, height, kind, *queue));
} }
@@ -781,7 +781,7 @@ hipError_t hipMemcpy2DAsync(void* dst, size_t dpitch, const void* src, size_t sp
queue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue(); queue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue();
} }
return ihipMemcpy2D(dst, dpitch, src, spitch, width, height, kind, *queue, true); HIP_RETURN(ihipMemcpy2D(dst, dpitch, src, spitch, width, height, kind, *queue, true));
} }
hipError_t hipMemcpy2DToArray(hipArray* dst, size_t wOffset, size_t hOffset, const void* src, hipError_t hipMemcpy2DToArray(hipArray* dst, size_t wOffset, size_t hOffset, const void* src,
@@ -789,7 +789,7 @@ hipError_t hipMemcpy2DToArray(hipArray* dst, size_t wOffset, size_t hOffset, con
HIP_INIT_API(dst, wOffset, hOffset, src, spitch, width, height, kind); HIP_INIT_API(dst, wOffset, hOffset, src, spitch, width, height, kind);
if (dst->data == nullptr) { if (dst->data == nullptr) {
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hip::syncStreams(); hip::syncStreams();
@@ -816,7 +816,7 @@ hipError_t hipMemcpy2DToArray(hipArray* dst, size_t wOffset, size_t hOffset, con
} }
if ((wOffset + width > (dpitch)) || width > spitch) { if ((wOffset + width > (dpitch)) || width > spitch) {
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
// Create buffer rectangle info structure // Create buffer rectangle info structure
@@ -836,7 +836,7 @@ hipError_t hipMemcpy2DToArray(hipArray* dst, size_t wOffset, size_t hOffset, con
if (!srcRect.create(sOrigin, region, spitch, src_slice_pitch) || if (!srcRect.create(sOrigin, region, spitch, src_slice_pitch) ||
!dstRect.create(dOrigin, region, dpitch, dst_slice_pitch)) { !dstRect.create(dOrigin, region, dpitch, dst_slice_pitch)) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
amd::Command* command = nullptr; amd::Command* command = nullptr;
@@ -855,7 +855,7 @@ hipError_t hipMemcpy2DToArray(hipArray* dst, size_t wOffset, size_t hOffset, con
void* pSrc = reinterpret_cast<void*>(reinterpret_cast<size_t>(src) + y * spitch); void* pSrc = reinterpret_cast<void*>(reinterpret_cast<size_t>(src) + y * spitch);
memcpy(pDst, pSrc, width); memcpy(pDst, pSrc, width);
} }
return hipSuccess; HIP_RETURN(hipSuccess);
} else if ((srcMemory == nullptr) && (dstMemory != nullptr)) { } else if ((srcMemory == nullptr) && (dstMemory != nullptr)) {
command = new amd::WriteMemoryCommand(*queue, CL_COMMAND_WRITE_BUFFER_RECT, waitList, command = new amd::WriteMemoryCommand(*queue, CL_COMMAND_WRITE_BUFFER_RECT, waitList,
*dstMemory->asBuffer(), dstStart, size, src, dstRect, srcRect); *dstMemory->asBuffer(), dstStart, size, src, dstRect, srcRect);
@@ -868,14 +868,14 @@ hipError_t hipMemcpy2DToArray(hipArray* dst, size_t wOffset, size_t hOffset, con
} }
if (command == nullptr) { if (command == nullptr) {
return hipErrorOutOfMemory; HIP_RETURN(hipErrorOutOfMemory);
} }
command->enqueue(); command->enqueue();
command->awaitCompletion(); command->awaitCompletion();
command->release(); command->release();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
@@ -909,14 +909,14 @@ hipError_t hipMemcpyToArray(hipArray* dstArray, size_t wOffset, size_t hOffset,
} }
if (command == nullptr) { if (command == nullptr) {
return hipErrorOutOfMemory; HIP_RETURN(hipErrorOutOfMemory);
} }
command->enqueue(); command->enqueue();
command->awaitCompletion(); command->awaitCompletion();
command->release(); command->release();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipMemcpyFromArray(void* dst, hipArray_const_t srcArray, size_t wOffset, size_t hOffset, hipError_t hipMemcpyFromArray(void* dst, hipArray_const_t srcArray, size_t wOffset, size_t hOffset,
@@ -949,14 +949,14 @@ hipError_t hipMemcpyFromArray(void* dst, hipArray_const_t srcArray, size_t wOffs
} }
if (command == nullptr) { if (command == nullptr) {
return hipErrorOutOfMemory; HIP_RETURN(hipErrorOutOfMemory);
} }
command->enqueue(); command->enqueue();
command->awaitCompletion(); command->awaitCompletion();
command->release(); command->release();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipMemcpyHtoA(hipArray* dstArray, size_t dstOffset, const void* srcHost, size_t count) { hipError_t hipMemcpyHtoA(hipArray* dstArray, size_t dstOffset, const void* srcHost, size_t count) {
@@ -986,14 +986,14 @@ hipError_t hipMemcpyHtoA(hipArray* dstArray, size_t dstOffset, const void* srcHo
} }
if (command == nullptr) { if (command == nullptr) {
return hipErrorOutOfMemory; HIP_RETURN(hipErrorOutOfMemory);
} }
command->enqueue(); command->enqueue();
command->awaitCompletion(); command->awaitCompletion();
command->release(); command->release();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipMemcpyAtoH(void* dst, hipArray* srcArray, size_t srcOffset, size_t count) { hipError_t hipMemcpyAtoH(void* dst, hipArray* srcArray, size_t srcOffset, size_t count) {
@@ -1023,14 +1023,14 @@ hipError_t hipMemcpyAtoH(void* dst, hipArray* srcArray, size_t srcOffset, size_t
} }
if (command == nullptr) { if (command == nullptr) {
return hipErrorOutOfMemory; HIP_RETURN(hipErrorOutOfMemory);
} }
command->enqueue(); command->enqueue();
command->awaitCompletion(); command->awaitCompletion();
command->release(); command->release();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipMemcpy3D(const struct hipMemcpy3DParms* p) { hipError_t hipMemcpy3D(const struct hipMemcpy3DParms* p) {
@@ -1108,7 +1108,7 @@ hipError_t hipMemcpy3D(const struct hipMemcpy3DParms* p) {
if (!srcRect.create(srcOrigin, region, srcPitchInBytes, src_slice_pitch) || if (!srcRect.create(srcOrigin, region, srcPitchInBytes, src_slice_pitch) ||
!dstRect.create(dstOrigin, region, dstPitchInbytes, dst_slice_pitch)) { !dstRect.create(dstOrigin, region, dstPitchInbytes, dst_slice_pitch)) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
hipMemcpyKind kind = p->kind; hipMemcpyKind kind = p->kind;
@@ -1123,7 +1123,7 @@ hipError_t hipMemcpy3D(const struct hipMemcpy3DParms* p) {
if (((srcMemory == nullptr) && (dstMemory == nullptr)) || if (((srcMemory == nullptr) && (dstMemory == nullptr)) ||
(kind == hipMemcpyHostToHost)) { (kind == hipMemcpyHostToHost)) {
memcpy(dstPtr, srcPtr, region[0] * region[1] * region[2]); memcpy(dstPtr, srcPtr, region[0] * region[1] * region[2]);
return hipSuccess; HIP_RETURN(hipSuccess);
} else if ((srcMemory == nullptr) && (dstMemory != nullptr)) { } else if ((srcMemory == nullptr) && (dstMemory != nullptr)) {
command = new amd::WriteMemoryCommand(*queue, CL_COMMAND_WRITE_BUFFER_RECT, waitList, command = new amd::WriteMemoryCommand(*queue, CL_COMMAND_WRITE_BUFFER_RECT, waitList,
*dstMemory->asBuffer(), srcStart, size, srcPtr, srcRect, dstRect); *dstMemory->asBuffer(), srcStart, size, srcPtr, srcRect, dstRect);
@@ -1137,14 +1137,14 @@ hipError_t hipMemcpy3D(const struct hipMemcpy3DParms* p) {
} }
if (command == nullptr) { if (command == nullptr) {
return hipErrorOutOfMemory; HIP_RETURN(hipErrorOutOfMemory);
} }
command->enqueue(); command->enqueue();
command->awaitCompletion(); command->awaitCompletion();
command->release(); command->release();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t ihipMemset2D(void* dst, size_t pitch, int value, size_t width, size_t height, hipError_t ihipMemset2D(void* dst, size_t pitch, int value, size_t width, size_t height,
@@ -1192,7 +1192,7 @@ hipError_t hipMemset2D(void* dst, size_t pitch, int value, size_t width, size_t
hip::syncStreams(); hip::syncStreams();
amd::HostQueue* queue = hip::getNullStream(); amd::HostQueue* queue = hip::getNullStream();
return ihipMemset2D(dst, pitch, value, width, height, *queue); HIP_RETURN(ihipMemset2D(dst, pitch, value, width, height, *queue));
} }
hipError_t hipMemset2DAsync(void* dst, size_t pitch, int value, hipError_t hipMemset2DAsync(void* dst, size_t pitch, int value,
@@ -1208,14 +1208,14 @@ hipError_t hipMemset2DAsync(void* dst, size_t pitch, int value,
queue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue(); queue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue();
} }
return ihipMemset2D(dst, pitch, value, width, height, *queue, true); HIP_RETURN(ihipMemset2D(dst, pitch, value, width, height, *queue, true));
} }
hipError_t hipMemsetD8(hipDeviceptr_t dst, unsigned char value, size_t sizeBytes) { hipError_t hipMemsetD8(hipDeviceptr_t dst, unsigned char value, size_t sizeBytes) {
HIP_INIT_API(dst, value, sizeBytes); HIP_INIT_API(dst, value, sizeBytes);
if (dst == nullptr) { if (dst == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
hip::syncStreams(); hip::syncStreams();
@@ -1232,7 +1232,7 @@ hipError_t hipMemsetD8(hipDeviceptr_t dst, unsigned char value, size_t sizeBytes
&value, sizeof(char), fillOffset, fillSize); &value, sizeof(char), fillOffset, fillSize);
if (command == nullptr) { if (command == nullptr) {
return hipErrorOutOfMemory; HIP_RETURN(hipErrorOutOfMemory);
} }
command->enqueue(); command->enqueue();
@@ -1243,7 +1243,7 @@ hipError_t hipMemsetD8(hipDeviceptr_t dst, unsigned char value, size_t sizeBytes
memset(dst, value, sizeBytes); memset(dst, value, sizeBytes);
} }
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipIpcGetMemHandle(hipIpcMemHandle_t* handle, void* devPtr) { hipError_t hipIpcGetMemHandle(hipIpcMemHandle_t* handle, void* devPtr) {
@@ -1251,7 +1251,7 @@ hipError_t hipIpcGetMemHandle(hipIpcMemHandle_t* handle, void* devPtr) {
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipIpcOpenMemHandle(void** devPtr, hipIpcMemHandle_t handle, unsigned int flags) { hipError_t hipIpcOpenMemHandle(void** devPtr, hipIpcMemHandle_t handle, unsigned int flags) {
@@ -1259,7 +1259,7 @@ hipError_t hipIpcOpenMemHandle(void** devPtr, hipIpcMemHandle_t handle, unsigned
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipIpcCloseMemHandle(void* devPtr) { hipError_t hipIpcCloseMemHandle(void* devPtr) {
@@ -1267,7 +1267,7 @@ hipError_t hipIpcCloseMemHandle(void* devPtr) {
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipChannelFormatDesc hipCreateChannelDesc(int x, int y, int z, int w, hipChannelFormatKind f) { hipChannelFormatDesc hipCreateChannelDesc(int x, int y, int z, int w, hipChannelFormatKind f) {
@@ -1287,11 +1287,11 @@ hipError_t hipHostGetDevicePointer(void** devicePointer, void* hostPointer, unsi
amd::Memory* memObj = getMemoryObject(hostPointer, offset); amd::Memory* memObj = getMemoryObject(hostPointer, offset);
if (!memObj) { if (!memObj) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
*devicePointer = reinterpret_cast<void*>(memObj->getDeviceMemory(*hip::getCurrentContext()->devices()[0])->virtualAddress() + offset); *devicePointer = reinterpret_cast<void*>(memObj->getDeviceMemory(*hip::getCurrentContext()->devices()[0])->virtualAddress() + offset);
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipPointerGetAttributes(hipPointerAttribute_t* attributes, const void* ptr) { hipError_t hipPointerGetAttributes(hipPointerAttribute_t* attributes, const void* ptr) {
@@ -1325,5 +1325,5 @@ hipError_t hipPointerGetAttributes(hipPointerAttribute_t* attributes, const void
attributes->allocationFlags = 0; attributes->allocationFlags = 0;
} }
return hipSuccess; HIP_RETURN(hipSuccess);
} }
projects/hip/api/hip/hip_module.cpp
+16 -16
View File
@@ -56,18 +56,18 @@ hipError_t hipModuleLoad(hipModule_t *module, const char *fname)
HIP_INIT_API(module, fname); HIP_INIT_API(module, fname);
if (!fname) { if (!fname) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
std::ifstream file{fname}; std::ifstream file{fname};
if (!file.is_open()) { if (!file.is_open()) {
return hipErrorFileNotFound; HIP_RETURN(hipErrorFileNotFound);
} }
std::vector<char> tmp{std::istreambuf_iterator<char>{file}, std::istreambuf_iterator<char>{}}; std::vector<char> tmp{std::istreambuf_iterator<char>{file}, std::istreambuf_iterator<char>{}};
return ihipModuleLoadData(module, tmp.data()); HIP_RETURN(ihipModuleLoadData(module, tmp.data()));
} }
@@ -76,21 +76,21 @@ hipError_t hipModuleUnload(hipModule_t hmod)
HIP_INIT_API(hmod); HIP_INIT_API(hmod);
if (hmod == nullptr) { if (hmod == nullptr) {
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
amd::Program* program = as_amd(reinterpret_cast<cl_program>(hmod)); amd::Program* program = as_amd(reinterpret_cast<cl_program>(hmod));
program->release(); program->release();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipModuleLoadData(hipModule_t *module, const void *image) hipError_t hipModuleLoadData(hipModule_t *module, const void *image)
{ {
HIP_INIT_API(module, image); HIP_INIT_API(module, image);
return ihipModuleLoadData(module, image); HIP_RETURN(ihipModuleLoadData(module, image));
} }
hipError_t ihipModuleLoadData(hipModule_t *module, const void *image) hipError_t ihipModuleLoadData(hipModule_t *module, const void *image)
@@ -118,24 +118,24 @@ hipError_t hipModuleGetFunction(hipFunction_t *hfunc, hipModule_t hmod, const ch
const amd::Symbol* symbol = program->findSymbol(name); const amd::Symbol* symbol = program->findSymbol(name);
if (!symbol) { if (!symbol) {
return hipErrorNotFound; HIP_RETURN(hipErrorNotFound);
} }
amd::Kernel* kernel = new amd::Kernel(*program, *symbol, name); amd::Kernel* kernel = new amd::Kernel(*program, *symbol, name);
if (!kernel) { if (!kernel) {
return hipErrorOutOfMemory; HIP_RETURN(hipErrorOutOfMemory);
} }
*hfunc = reinterpret_cast<hipFunction_t>(as_cl(kernel)); *hfunc = reinterpret_cast<hipFunction_t>(as_cl(kernel));
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipFuncGetAttributes(hipFuncAttributes* attr, const void* func) hipError_t hipFuncGetAttributes(hipFuncAttributes* attr, const void* func)
{ {
HIP_INIT_API(attr, func); HIP_INIT_API(attr, func);
return hipErrorInvalidDeviceFunction; HIP_RETURN(hipErrorInvalidDeviceFunction);
} }
@@ -242,9 +242,9 @@ hipError_t hipModuleLaunchKernel(hipFunction_t f,
uint32_t sharedMemBytes, hipStream_t hStream, uint32_t sharedMemBytes, hipStream_t hStream,
void **kernelParams, void **extra) void **kernelParams, void **extra)
{ {
return ihipModuleLaunchKernel(f, gridDimX * blockDimX, gridDimY * blockDimY, gridDimZ * blockDimZ, HIP_RETURN(ihipModuleLaunchKernel(f, gridDimX * blockDimX, gridDimY * blockDimY, gridDimZ * blockDimZ,
blockDimX, blockDimY, blockDimZ, blockDimX, blockDimY, blockDimZ,
sharedMemBytes, hStream, kernelParams, extra, nullptr, nullptr); sharedMemBytes, hStream, kernelParams, extra, nullptr, nullptr));
} }
hipError_t hipHccModuleLaunchKernel(hipFunction_t f, uint32_t gridDimX, hipError_t hipHccModuleLaunchKernel(hipFunction_t f, uint32_t gridDimX,
@@ -255,8 +255,8 @@ hipError_t hipHccModuleLaunchKernel(hipFunction_t f, uint32_t gridDimX,
hipEvent_t startEvent, hipEvent_t startEvent,
hipEvent_t stopEvent) hipEvent_t stopEvent)
{ {
return ihipModuleLaunchKernel(f, gridDimX, gridDimY, gridDimZ, blockDimX, blockDimY, blockDimZ, HIP_RETURN(ihipModuleLaunchKernel(f, gridDimX, gridDimY, gridDimZ, blockDimX, blockDimY, blockDimZ,
sharedMemBytes, hStream, kernelParams, extra, startEvent, stopEvent); sharedMemBytes, hStream, kernelParams, extra, startEvent, stopEvent));
} }
hipError_t hipModuleLaunchKernelExt(hipFunction_t f, uint32_t gridDimX, hipError_t hipModuleLaunchKernelExt(hipFunction_t f, uint32_t gridDimX,
@@ -267,8 +267,8 @@ hipError_t hipModuleLaunchKernelExt(hipFunction_t f, uint32_t gridDimX,
hipEvent_t startEvent, hipEvent_t startEvent,
hipEvent_t stopEvent) hipEvent_t stopEvent)
{ {
return ihipModuleLaunchKernel(f, gridDimX, gridDimY, gridDimZ, blockDimX, blockDimY, blockDimZ, HIP_RETURN(ihipModuleLaunchKernel(f, gridDimX, gridDimY, gridDimZ, blockDimX, blockDimY, blockDimZ,
sharedMemBytes, hStream, kernelParams, extra, startEvent, stopEvent); sharedMemBytes, hStream, kernelParams, extra, startEvent, stopEvent));
} }
projects/hip/api/hip/hip_peer.cpp
+10 -10
View File
@@ -29,7 +29,7 @@ hipError_t hipDeviceCanAccessPeer(int* canAccessPeer, hipCtx_t thisCtx, hipCtx_t
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipMemcpyPeer(void* dst, hipCtx_t dstCtx, const void* src, hipCtx_t srcCtx, hipError_t hipMemcpyPeer(void* dst, hipCtx_t dstCtx, const void* src, hipCtx_t srcCtx,
@@ -38,7 +38,7 @@ hipError_t hipMemcpyPeer(void* dst, hipCtx_t dstCtx, const void* src, hipCtx_t s
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipMemcpyPeerAsync(void* dst, hipCtx_t dstDevice, const void* src, hipCtx_t srcDevice, hipError_t hipMemcpyPeerAsync(void* dst, hipCtx_t dstDevice, const void* src, hipCtx_t srcDevice,
@@ -47,7 +47,7 @@ hipError_t hipMemcpyPeerAsync(void* dst, hipCtx_t dstDevice, const void* src, hi
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipDeviceCanAccessPeer(int* canAccessPeer, int deviceId, int peerDeviceId) { hipError_t hipDeviceCanAccessPeer(int* canAccessPeer, int deviceId, int peerDeviceId) {
@@ -55,7 +55,7 @@ hipError_t hipDeviceCanAccessPeer(int* canAccessPeer, int deviceId, int peerDevi
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipDeviceDisablePeerAccess(int peerDeviceId) { hipError_t hipDeviceDisablePeerAccess(int peerDeviceId) {
@@ -63,7 +63,7 @@ hipError_t hipDeviceDisablePeerAccess(int peerDeviceId) {
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipDeviceEnablePeerAccess(int peerDeviceId, unsigned int flags) { hipError_t hipDeviceEnablePeerAccess(int peerDeviceId, unsigned int flags) {
@@ -71,7 +71,7 @@ hipError_t hipDeviceEnablePeerAccess(int peerDeviceId, unsigned int flags) {
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipMemcpyPeer(void* dst, int dstDevice, const void* src, int srcDevice, hipError_t hipMemcpyPeer(void* dst, int dstDevice, const void* src, int srcDevice,
@@ -80,7 +80,7 @@ hipError_t hipMemcpyPeer(void* dst, int dstDevice, const void* src, int srcDevic
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipMemcpyPeerAsync(void* dst, int dstDevice, const void* src, int srcDevice, hipError_t hipMemcpyPeerAsync(void* dst, int dstDevice, const void* src, int srcDevice,
@@ -89,7 +89,7 @@ hipError_t hipMemcpyPeerAsync(void* dst, int dstDevice, const void* src, int src
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipCtxEnablePeerAccess(hipCtx_t peerCtx, unsigned int flags) { hipError_t hipCtxEnablePeerAccess(hipCtx_t peerCtx, unsigned int flags) {
@@ -97,7 +97,7 @@ hipError_t hipCtxEnablePeerAccess(hipCtx_t peerCtx, unsigned int flags) {
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipCtxDisablePeerAccess(hipCtx_t peerCtx) { hipError_t hipCtxDisablePeerAccess(hipCtx_t peerCtx) {
@@ -105,5 +105,5 @@ hipError_t hipCtxDisablePeerAccess(hipCtx_t peerCtx) {
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
projects/hip/api/hip/hip_platform.cpp
+9 -8
View File
@@ -252,7 +252,7 @@ extern "C" hipError_t hipConfigureCall(
PlatformState::instance().configureCall(gridDim, blockDim, sharedMem, stream); PlatformState::instance().configureCall(gridDim, blockDim, sharedMem, stream);
return hipSuccess; HIP_RETURN(hipSuccess);
} }
extern "C" hipError_t hipSetupArgument( extern "C" hipError_t hipSetupArgument(
@@ -264,7 +264,7 @@ extern "C" hipError_t hipSetupArgument(
PlatformState::instance().setupArgument(arg, size, offset); PlatformState::instance().setupArgument(arg, size, offset);
return hipSuccess; HIP_RETURN(hipSuccess);
} }
extern "C" hipError_t hipLaunchByPtr(const void *hostFunction) extern "C" hipError_t hipLaunchByPtr(const void *hostFunction)
@@ -272,8 +272,9 @@ extern "C" hipError_t hipLaunchByPtr(const void *hostFunction)
HIP_INIT_API(hostFunction); HIP_INIT_API(hostFunction);
hipFunction_t func = PlatformState::instance().getFunc(hostFunction); hipFunction_t func = PlatformState::instance().getFunc(hostFunction);
if (func == nullptr) if (func == nullptr) {
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
}
ihipExec_t exec; ihipExec_t exec;
PlatformState::instance().popExec(exec); PlatformState::instance().popExec(exec);
@@ -284,10 +285,10 @@ extern "C" hipError_t hipLaunchByPtr(const void *hostFunction)
HIP_LAUNCH_PARAM_END HIP_LAUNCH_PARAM_END
}; };
return hipModuleLaunchKernel(func, HIP_RETURN(hipModuleLaunchKernel(func,
exec.gridDim_.x, exec.gridDim_.y, exec.gridDim_.z, exec.gridDim_.x, exec.gridDim_.y, exec.gridDim_.z,
exec.blockDim_.x, exec.blockDim_.y, exec.blockDim_.z, exec.blockDim_.x, exec.blockDim_.y, exec.blockDim_.z,
exec.sharedMem_, exec.hStream_, nullptr, extra); exec.sharedMem_, exec.hStream_, nullptr, extra));
} }
#if defined(ATI_OS_LINUX) #if defined(ATI_OS_LINUX)
@@ -514,7 +515,7 @@ static inline std::uint32_t f32_as_u32(float f) { union { float f; std::uint32_t
static inline float u32_as_f32(std::uint32_t u) { union { float f; std::uint32_t u; } v; v.u = u; return v.f; } static inline float u32_as_f32(std::uint32_t u) { union { float f; std::uint32_t u; } v; v.u = u; return v.f; }
static inline int clamp_int(int i, int l, int h) { return std::min(std::max(i, l), h); } static inline int clamp_int(int i, int l, int h) { return std::min(std::max(i, l), h); }
// half � float, the f16 is in the low 16 bits of the input argument �a� // half float, the f16 is in the low 16 bits of the input argument
static inline float __convert_half_to_float(std::uint32_t a) noexcept { static inline float __convert_half_to_float(std::uint32_t a) noexcept {
std::uint32_t u = ((a << 13) + 0x70000000U) & 0x8fffe000U; std::uint32_t u = ((a << 13) + 0x70000000U) & 0x8fffe000U;
std::uint32_t v = f32_as_u32(u32_as_f32(u) * 0x1.0p+112f) + 0x38000000U; std::uint32_t v = f32_as_u32(u32_as_f32(u) * 0x1.0p+112f) + 0x38000000U;
@@ -522,7 +523,7 @@ static inline float __convert_half_to_float(std::uint32_t a) noexcept {
return u32_as_f32(u) * 0x1.0p-112f; return u32_as_f32(u) * 0x1.0p-112f;
} }
// float � half with nearest even rounding // float half with nearest even rounding
// The lower 16 bits of the result is the bit pattern for the f16 // The lower 16 bits of the result is the bit pattern for the f16
static inline std::uint32_t __convert_float_to_half(float a) noexcept { static inline std::uint32_t __convert_float_to_half(float a) noexcept {
std::uint32_t u = f32_as_u32(a); std::uint32_t u = f32_as_u32(a);
projects/hip/api/hip/hip_profile.cpp
+2 -2
View File
@@ -29,7 +29,7 @@ hipError_t hipProfilerStart() {
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
@@ -38,5 +38,5 @@ hipError_t hipProfilerStop() {
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
projects/hip/api/hip/hip_stream.cpp
+14 -14
View File
@@ -68,13 +68,13 @@ static hipError_t ihipStreamCreateWithFlags(hipStream_t *stream, unsigned int fl
hipError_t hipStreamCreateWithFlags(hipStream_t *stream, unsigned int flags) { hipError_t hipStreamCreateWithFlags(hipStream_t *stream, unsigned int flags) {
HIP_INIT_API(stream, flags); HIP_INIT_API(stream, flags);
return ihipStreamCreateWithFlags(stream, flags); HIP_RETURN(ihipStreamCreateWithFlags(stream, flags));
} }
hipError_t hipStreamCreate(hipStream_t *stream) { hipError_t hipStreamCreate(hipStream_t *stream) {
HIP_INIT_API(stream); HIP_INIT_API(stream);
return ihipStreamCreateWithFlags(stream, hipStreamDefault); HIP_RETURN(ihipStreamCreateWithFlags(stream, hipStreamDefault));
} }
hipError_t hipStreamGetFlags(hipStream_t stream, unsigned int *flags) { hipError_t hipStreamGetFlags(hipStream_t stream, unsigned int *flags) {
@@ -86,10 +86,10 @@ hipError_t hipStreamGetFlags(hipStream_t stream, unsigned int *flags) {
if(flags != nullptr) { if(flags != nullptr) {
*flags = (it == streamSet.end()) ? hipStreamNonBlocking : hipStreamDefault; *flags = (it == streamSet.end()) ? hipStreamNonBlocking : hipStreamDefault;
} else { } else {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipStreamSynchronize(hipStream_t stream) { hipError_t hipStreamSynchronize(hipStream_t stream) {
@@ -108,19 +108,19 @@ hipError_t hipStreamSynchronize(hipStream_t stream) {
} }
if (hostQueue == nullptr) { if (hostQueue == nullptr) {
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hostQueue->finish(); hostQueue->finish();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipStreamDestroy(hipStream_t stream) { hipError_t hipStreamDestroy(hipStream_t stream) {
HIP_INIT_API(stream); HIP_INIT_API(stream);
if (stream == nullptr) { if (stream == nullptr) {
return hipErrorInvalidResourceHandle; HIP_RETURN(hipErrorInvalidResourceHandle);
} }
amd::ScopedLock lock(streamSetLock); amd::ScopedLock lock(streamSetLock);
@@ -129,14 +129,14 @@ hipError_t hipStreamDestroy(hipStream_t stream) {
hostQueue->release(); hostQueue->release();
streamSet.erase(hostQueue); streamSet.erase(hostQueue);
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipStreamWaitEvent(hipStream_t stream, hipEvent_t event, unsigned int flags) { hipError_t hipStreamWaitEvent(hipStream_t stream, hipEvent_t event, unsigned int flags) {
HIP_INIT_API(stream, event, flags); HIP_INIT_API(stream, event, flags);
if (stream == nullptr || event == nullptr) { if (stream == nullptr || event == nullptr) {
return hipErrorInvalidResourceHandle; HIP_RETURN(hipErrorInvalidResourceHandle);
} }
amd::HostQueue* hostQueue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue(); amd::HostQueue* hostQueue = as_amd(reinterpret_cast<cl_command_queue>(stream))->asHostQueue();
@@ -146,17 +146,17 @@ hipError_t hipStreamWaitEvent(hipStream_t stream, hipEvent_t event, unsigned int
amd::Command::EventWaitList eventWaitList; amd::Command::EventWaitList eventWaitList;
cl_int err = amd::clSetEventWaitList(eventWaitList, *hostQueue, 1, &clEvent); cl_int err = amd::clSetEventWaitList(eventWaitList, *hostQueue, 1, &clEvent);
if (err != CL_SUCCESS) { if (err != CL_SUCCESS) {
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
amd::Command* command = new amd::Marker(*hostQueue, true, eventWaitList); amd::Command* command = new amd::Marker(*hostQueue, true, eventWaitList);
if (command == NULL) { if (command == NULL) {
return hipErrorOutOfMemory; HIP_RETURN(hipErrorOutOfMemory);
} }
command->enqueue(); command->enqueue();
command->release(); command->release();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipStreamQuery(hipStream_t stream) { hipError_t hipStreamQuery(hipStream_t stream) {
@@ -164,7 +164,7 @@ hipError_t hipStreamQuery(hipStream_t stream) {
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipStreamAddCallback(hipStream_t stream, hipStreamCallback_t callback, void* userData, hipError_t hipStreamAddCallback(hipStream_t stream, hipStreamCallback_t callback, void* userData,
@@ -173,7 +173,7 @@ hipError_t hipStreamAddCallback(hipStream_t stream, hipStreamCallback_t callback
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
projects/hip/api/hip/hip_surface.cpp
+2 -2
View File
@@ -36,7 +36,7 @@ hipError_t hipCreateSurfaceObject(hipSurfaceObject_t* pSurfObject,
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
@@ -45,5 +45,5 @@ hipError_t hipDestroySurfaceObject(hipSurfaceObject_t surfaceObject) {
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
projects/hip/api/hip/hip_texture.cpp
+29 -29
View File
@@ -145,7 +145,7 @@ hipError_t hipCreateTextureObject(hipTextureObject_t* pTexObject, const hipResou
amd::Device* device = hip::getCurrentContext()->devices()[0]; amd::Device* device = hip::getCurrentContext()->devices()[0];
if (!device->info().imageSupport_) { if (!device->info().imageSupport_) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
amd::Image* image = nullptr; amd::Image* image = nullptr;
@@ -203,11 +203,11 @@ hipError_t hipCreateTextureObject(hipTextureObject_t* pTexObject, const hipResou
pResDesc->res.pitch2D.width, pResDesc->res.pitch2D.height, 1, pResDesc->res.pitch2D.width, pResDesc->res.pitch2D.height, 1,
pResDesc->res.pitch2D.pitchInBytes, 0); pResDesc->res.pitch2D.pitchInBytes, 0);
break; break;
default: return hipErrorInvalidValue; default: HIP_RETURN(hipErrorInvalidValue);
} }
*pTexObject = reinterpret_cast<hipTextureObject_t>(as_cl(image)); *pTexObject = reinterpret_cast<hipTextureObject_t>(as_cl(image));
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipDestroyTextureObject(hipTextureObject_t textureObject) { hipError_t hipDestroyTextureObject(hipTextureObject_t textureObject) {
@@ -215,7 +215,7 @@ hipError_t hipDestroyTextureObject(hipTextureObject_t textureObject) {
as_amd(reinterpret_cast<cl_mem>(textureObject))->release(); as_amd(reinterpret_cast<cl_mem>(textureObject))->release();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipGetTextureObjectResourceDesc(hipResourceDesc* pResDesc, hipError_t hipGetTextureObjectResourceDesc(hipResourceDesc* pResDesc,
@@ -224,7 +224,7 @@ hipError_t hipGetTextureObjectResourceDesc(hipResourceDesc* pResDesc,
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipGetTextureObjectResourceViewDesc(hipResourceViewDesc* pResViewDesc, hipError_t hipGetTextureObjectResourceViewDesc(hipResourceViewDesc* pResViewDesc,
@@ -233,7 +233,7 @@ hipError_t hipGetTextureObjectResourceViewDesc(hipResourceViewDesc* pResViewDesc
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipGetTextureObjectTextureDesc(hipTextureDesc* pTexDesc, hipError_t hipGetTextureObjectTextureDesc(hipTextureDesc* pTexDesc,
@@ -242,7 +242,7 @@ hipError_t hipGetTextureObjectTextureDesc(hipTextureDesc* pTexDesc,
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t ihipBindTexture(cl_mem_object_type type, hipError_t ihipBindTexture(cl_mem_object_type type,
@@ -282,14 +282,14 @@ hipError_t hipBindTexture(size_t* offset, textureReference* tex, const void* dev
HIP_INIT_API(offset, tex, devPtr, desc, size); HIP_INIT_API(offset, tex, devPtr, desc, size);
if (desc == nullptr) { if (desc == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
cl_image_format image_format; cl_image_format image_format;
getChannelOrderAndType(*desc, hipReadModeElementType, getChannelOrderAndType(*desc, hipReadModeElementType,
&image_format.image_channel_order, &image_format.image_channel_data_type); &image_format.image_channel_order, &image_format.image_channel_data_type);
const amd::Image::Format imageFormat(image_format); const amd::Image::Format imageFormat(image_format);
return ihipBindTexture(CL_MEM_OBJECT_IMAGE1D, offset, tex, devPtr, desc, size / imageFormat.getElementSize(), 1, size); HIP_RETURN(ihipBindTexture(CL_MEM_OBJECT_IMAGE1D, offset, tex, devPtr, desc, size / imageFormat.getElementSize(), 1, size));
} }
hipError_t hipBindTexture2D(size_t* offset, textureReference* tex, const void* devPtr, hipError_t hipBindTexture2D(size_t* offset, textureReference* tex, const void* devPtr,
@@ -297,7 +297,7 @@ hipError_t hipBindTexture2D(size_t* offset, textureReference* tex, const void* d
size_t pitch) { size_t pitch) {
HIP_INIT_API(offset, tex, devPtr, desc, width, height, pitch); HIP_INIT_API(offset, tex, devPtr, desc, width, height, pitch);
return ihipBindTexture(CL_MEM_OBJECT_IMAGE2D, offset, tex, devPtr, desc, width, height, pitch); HIP_RETURN(ihipBindTexture(CL_MEM_OBJECT_IMAGE2D, offset, tex, devPtr, desc, width, height, pitch));
} }
hipError_t hipBindTextureToArray(textureReference* tex, hipArray_const_t array, hipError_t hipBindTextureToArray(textureReference* tex, hipArray_const_t array,
@@ -306,7 +306,7 @@ hipError_t hipBindTextureToArray(textureReference* tex, hipArray_const_t array,
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t ihipBindTextureToArrayImpl(int dim, enum hipTextureReadMode readMode, hipError_t ihipBindTextureToArrayImpl(int dim, enum hipTextureReadMode readMode,
@@ -325,7 +325,7 @@ hipError_t hipBindTextureToMipmappedArray(textureReference* tex,
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipUnbindTexture(const textureReference* tex) { hipError_t hipUnbindTexture(const textureReference* tex) {
@@ -333,7 +333,7 @@ hipError_t hipUnbindTexture(const textureReference* tex) {
as_amd(reinterpret_cast<cl_mem>(tex->textureObject))->release(); as_amd(reinterpret_cast<cl_mem>(tex->textureObject))->release();
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipGetChannelDesc(hipChannelFormatDesc* desc, hipArray_const_t array) { hipError_t hipGetChannelDesc(hipChannelFormatDesc* desc, hipArray_const_t array) {
@@ -341,7 +341,7 @@ hipError_t hipGetChannelDesc(hipChannelFormatDesc* desc, hipArray_const_t array)
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipGetTextureAlignmentOffset(size_t* offset, const textureReference* tex) { hipError_t hipGetTextureAlignmentOffset(size_t* offset, const textureReference* tex) {
@@ -349,7 +349,7 @@ hipError_t hipGetTextureAlignmentOffset(size_t* offset, const textureReference*
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipGetTextureReference(const textureReference** tex, const void* symbol) { hipError_t hipGetTextureReference(const textureReference** tex, const void* symbol) {
@@ -357,56 +357,56 @@ hipError_t hipGetTextureReference(const textureReference** tex, const void* symb
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipTexRefSetFormat(textureReference* tex, hipArray_Format fmt, int NumPackedComponents) { hipError_t hipTexRefSetFormat(textureReference* tex, hipArray_Format fmt, int NumPackedComponents) {
HIP_INIT_API(tex, fmt, NumPackedComponents); HIP_INIT_API(tex, fmt, NumPackedComponents);
if (tex == nullptr) { if (tex == nullptr) {
return hipErrorInvalidImage; HIP_RETURN(hipErrorInvalidImage);
} }
tex->format = fmt; tex->format = fmt;
tex->numChannels = NumPackedComponents; tex->numChannels = NumPackedComponents;
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipTexRefSetFlags(textureReference* tex, unsigned int flags) { hipError_t hipTexRefSetFlags(textureReference* tex, unsigned int flags) {
HIP_INIT_API(tex, flags); HIP_INIT_API(tex, flags);
if (tex == nullptr) { if (tex == nullptr) {
return hipErrorInvalidImage; HIP_RETURN(hipErrorInvalidImage);
} }
tex->normalized = flags; tex->normalized = flags;
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipTexRefSetFilterMode(textureReference* tex, hipTextureFilterMode fm) { hipError_t hipTexRefSetFilterMode(textureReference* tex, hipTextureFilterMode fm) {
HIP_INIT_API(tex, fm); HIP_INIT_API(tex, fm);
if (tex == nullptr) { if (tex == nullptr) {
return hipErrorInvalidImage; HIP_RETURN(hipErrorInvalidImage);
} }
tex->filterMode = fm; tex->filterMode = fm;
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipTexRefSetAddressMode(textureReference* tex, int dim, hipTextureAddressMode am) { hipError_t hipTexRefSetAddressMode(textureReference* tex, int dim, hipTextureAddressMode am) {
HIP_INIT_API(tex, dim, am); HIP_INIT_API(tex, dim, am);
if (tex == nullptr) { if (tex == nullptr) {
return hipErrorInvalidImage; HIP_RETURN(hipErrorInvalidImage);
} }
tex->addressMode[dim] = am; tex->addressMode[dim] = am;
return hipSuccess; HIP_RETURN(hipSuccess);
} }
hipError_t hipTexRefSetArray(textureReference* tex, hipArray_const_t array, unsigned int flags) { hipError_t hipTexRefSetArray(textureReference* tex, hipArray_const_t array, unsigned int flags) {
@@ -414,7 +414,7 @@ hipError_t hipTexRefSetArray(textureReference* tex, hipArray_const_t array, unsi
assert(0 && "Unimplemented"); assert(0 && "Unimplemented");
return hipErrorUnknown; HIP_RETURN(hipErrorUnknown);
} }
hipError_t hipTexRefSetAddress(size_t* offset, textureReference* tex, hipDeviceptr_t devPtr, hipError_t hipTexRefSetAddress(size_t* offset, textureReference* tex, hipDeviceptr_t devPtr,
@@ -422,7 +422,7 @@ hipError_t hipTexRefSetAddress(size_t* offset, textureReference* tex, hipDevicep
HIP_INIT_API(offset, tex, devPtr, size); HIP_INIT_API(offset, tex, devPtr, size);
if (tex == nullptr) { if (tex == nullptr) {
return hipErrorInvalidImage; HIP_RETURN(hipErrorInvalidImage);
} }
cl_image_format image_format; cl_image_format image_format;
@@ -430,7 +430,7 @@ hipError_t hipTexRefSetAddress(size_t* offset, textureReference* tex, hipDevicep
&image_format.image_channel_order, &image_format.image_channel_data_type); &image_format.image_channel_order, &image_format.image_channel_data_type);
const amd::Image::Format imageFormat(image_format); const amd::Image::Format imageFormat(image_format);
return ihipBindTexture(CL_MEM_OBJECT_IMAGE1D, offset, tex, devPtr, nullptr, size / imageFormat.getElementSize(), 1, size); HIP_RETURN(ihipBindTexture(CL_MEM_OBJECT_IMAGE1D, offset, tex, devPtr, nullptr, size / imageFormat.getElementSize(), 1, size));
} }
hipError_t hipTexRefSetAddress2D(textureReference* tex, const HIP_ARRAY_DESCRIPTOR* desc, hipError_t hipTexRefSetAddress2D(textureReference* tex, const HIP_ARRAY_DESCRIPTOR* desc,
@@ -438,9 +438,9 @@ hipError_t hipTexRefSetAddress2D(textureReference* tex, const HIP_ARRAY_DESCRIPT
HIP_INIT_API(tex, desc, devPtr, pitch); HIP_INIT_API(tex, desc, devPtr, pitch);
if (desc == nullptr) { if (desc == nullptr) {
return hipErrorInvalidValue; HIP_RETURN(hipErrorInvalidValue);
} }
size_t offset; size_t offset;
return ihipBindTexture(CL_MEM_OBJECT_IMAGE2D, &offset, tex, devPtr, nullptr, desc->width, desc->height, pitch); HIP_RETURN(ihipBindTexture(CL_MEM_OBJECT_IMAGE2D, &offset, tex, devPtr, nullptr, desc->width, desc->height, pitch));
} }