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/clr commit: 2ec2cfd3c1]
This commit is contained in:
foreman
2018-08-14 18:54:13 -04:00
parent 5c6917d511
commit c1bb44fa73
14 changed files with 396 additions and 265 deletions
Download Patch File Download Diff File Expand all files Collapse all files
projects/clr/hipamd/api/hip/hip_context.cpp
+21 -20
View File
@@ -31,6 +31,7 @@ namespace hip {
thread_local amd::Context* g_context = nullptr;
thread_local std::stack<amd::Context*> g_ctxtStack;
thread_local hipError_t g_lastError = hipSuccess;
std::once_flag g_ihipInitialized;
std::map<amd::Context*, amd::HostQueue*> g_nullStreams;
@@ -85,14 +86,14 @@ using namespace hip;
hipError_t hipInit(unsigned int flags) {
HIP_INIT_API(flags);
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipCtxCreate(hipCtx_t *ctx, unsigned int flags, hipDevice_t device) {
HIP_INIT_API(ctx, flags, device);
if (static_cast<size_t>(device) >= g_devices.size()) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
*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_ctxtStack.push(g_devices[device]);
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipCtxSetCurrent(hipCtx_t ctx) {
@@ -119,7 +120,7 @@ hipError_t hipCtxSetCurrent(hipCtx_t ctx) {
g_ctxtStack.push(hip::getCurrentContext());
}
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipCtxGetCurrent(hipCtx_t* ctx) {
@@ -127,19 +128,19 @@ hipError_t hipCtxGetCurrent(hipCtx_t* ctx) {
*ctx = reinterpret_cast<hipCtx_t>(hip::getCurrentContext());
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipRuntimeGetVersion(int *runtimeVersion) {
HIP_INIT_API(runtimeVersion);
if (!runtimeVersion) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
*runtimeVersion = AMD_PLATFORM_BUILD_NUMBER;
return hipSuccess;
HIP_RETURN(hipSuccess);
}
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));
if (amdContext == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
// 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) {
@@ -171,17 +172,17 @@ hipError_t hipCtxPopCurrent(hipCtx_t* ctx) {
amd::Context* amdContext = reinterpret_cast<amd::Context*>(as_amd(ctx));
if (amdContext == nullptr) {
return hipErrorInvalidContext;
HIP_RETURN(hipErrorInvalidContext);
}
if (!g_ctxtStack.empty()) {
amdContext = g_ctxtStack.top();
g_ctxtStack.pop();
} else {
return hipErrorInvalidContext;
HIP_RETURN(hipErrorInvalidContext);
}
return hipSuccess;
HIP_RETURN(hipSuccess);
}
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));
if (amdContext == nullptr) {
return hipErrorInvalidContext;
HIP_RETURN(hipErrorInvalidContext);
}
hip::g_context = amdContext;
g_ctxtStack.push(hip::getCurrentContext());
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipDriverGetVersion(int* driverVersion) {
@@ -205,13 +206,13 @@ hipError_t hipDriverGetVersion(int* driverVersion) {
const auto& info = deviceHandle->info();
if (driverVersion) {
*driverVersion = AMD_PLATFORM_BUILD_NUMBER * 100 +
AMD_PLATFORM_REVISION_NUMBER;
*driverVersion = AMD_PLATFORM_BUILD_NUMBER * 100 +
AMD_PLATFORM_REVISION_NUMBER;
} else {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
return hipSuccess;;
HIP_RETURN(hipSuccess);
}
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++) {
if (g_devices[i] == hip::getCurrentContext()) {
*device = static_cast<hipDevice_t>(i);
return hipSuccess;
HIP_RETURN(hipSuccess);
}
}
} else {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
return hipErrorUnknown;
projects/clr/hipamd/api/hip/hip_device.cpp
+18 -18
View File
@@ -30,10 +30,10 @@ hipError_t hipDeviceGet(hipDevice_t *device, int deviceId) {
if (device != nullptr) {
*device = deviceId;
} else {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
return hipSuccess;
HIP_RETURN(hipSuccess);
};
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.
return hipSuccess;
HIP_RETURN(hipSuccess);
}
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);
if (device < 0 || static_cast<size_t>(device) >= g_devices.size()) {
return hipErrorInvalidDevice;
HIP_RETURN(hipErrorInvalidDevice);
}
if (bytes == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
auto* deviceHandle = g_devices[device]->devices()[0];
@@ -62,7 +62,7 @@ hipError_t hipDeviceTotalMem (size_t *bytes, hipDevice_t device) {
*bytes = info.globalMemSize_;
return hipSuccess;
HIP_RETURN(hipSuccess);
}
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);
if (device < 0 || static_cast<size_t>(device) >= g_devices.size()) {
return hipErrorInvalidDevice;
HIP_RETURN(hipErrorInvalidDevice);
}
if (major == nullptr || minor == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
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;
*minor = info.gfxipVersion_ % 100;
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipDeviceGetCount(int* count) {
HIP_INIT_API(count);
return ihipDeviceGetCount(count);
HIP_RETURN(ihipDeviceGetCount(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);
if (device < 0 || static_cast<size_t>(device) >= g_devices.size()) {
return hipErrorInvalidDevice;
HIP_RETURN(hipErrorInvalidDevice);
}
if (name == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
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;
::strncpy(name, info.boardName_, len);
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipGetDeviceProperties ( hipDeviceProp_t* props, hipDevice_t device ) {
HIP_INIT_API(props, device);
if (props == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
if (unsigned(device) >= g_devices.size()) {
return hipErrorInvalidDevice;
HIP_RETURN(hipErrorInvalidDevice);
}
auto* deviceHandle = g_devices[device]->devices()[0];
@@ -188,7 +188,7 @@ hipError_t hipGetDeviceProperties ( hipDeviceProp_t* props, hipDevice_t device )
deviceProps.gcnArch = info.gfxipVersion_;
*props = deviceProps;
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipHccGetAccelerator(int deviceId, hc::accelerator* acc) {
@@ -196,7 +196,7 @@ hipError_t hipHccGetAccelerator(int deviceId, hc::accelerator* acc) {
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
projects/clr/hipamd/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);
if (device == nullptr || properties == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
*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) {
@@ -143,19 +143,19 @@ hipError_t hipDeviceGetAttribute(int* pi, hipDeviceAttribute_t attr, int device)
HIP_INIT_API(pi, attr, device);
if (pi == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
int count = 0;
ihipDeviceGetCount(&count);
if (device < 0 || device >= count) {
return hipErrorInvalidDevice;
HIP_RETURN(hipErrorInvalidDevice);
}
//FIXME: should we cache the props, or just select from deviceHandle->info_?
hipDeviceProp_t prop = {0};
hipError_t err = hipGetDeviceProperties(&prop, device);
if (err != hipSuccess) return err;
if (err != hipSuccess) HIP_RETURN(err);
switch (attr) {
case hipDeviceAttributeMaxThreadsPerBlock:
@@ -234,10 +234,10 @@ hipError_t hipDeviceGetAttribute(int* pi, hipDeviceAttribute_t attr, int device)
*pi = prop.isMultiGpuBoard;
break;
default:
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
return hipSuccess;
HIP_RETURN(hipSuccess);
}
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);
if (device == nullptr || pciBusIdstr == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
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 ) {
HIP_INIT_API(cacheConfig);
if(cacheConfig == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
*cacheConfig = hipFuncCache_t();
return hipSuccess;
HIP_RETURN(hipSuccess);
}
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);
if(pValue == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
if(limit == hipLimitMallocHeapSize) {
hipDeviceProp_t prop;
hipGetDeviceProperties(&prop, 0);
*pValue = prop.totalGlobalMem;
return hipSuccess;
HIP_RETURN(hipSuccess);
} else {
return hipErrorUnsupportedLimit;
HIP_RETURN(hipErrorUnsupportedLimit);
}
}
/**
hipError_t hipDeviceGetP2PAttribute ( int* value, hipDeviceP2PAttr attr, int srcDevice, int dstDevice ) {
assert(0);
return hipSuccess;
HIP_RETURN(hipSuccess);
}
**/
@@ -315,11 +315,11 @@ hipError_t hipDeviceGetPCIBusId ( char* pciBusId, int len, int device ) {
int count;
ihipDeviceGetCount(&count);
if (device < 0 || device > count) {
return hipErrorInvalidDevice;
HIP_RETURN(hipErrorInvalidDevice);
}
if (pciBusId == nullptr || len < 0) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
hipDeviceProp_t prop;
@@ -330,7 +330,7 @@ hipError_t hipDeviceGetPCIBusId ( char* pciBusId, int len, int device ) {
prop.pciBusID,
prop.pciDeviceID);
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipDeviceGetSharedMemConfig ( hipSharedMemConfig * pConfig ) {
@@ -338,12 +338,12 @@ hipError_t hipDeviceGetSharedMemConfig ( hipSharedMemConfig * pConfig ) {
*pConfig = hipSharedMemBankSizeFourByte;
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipDeviceGetStreamPriorityRange ( int* leastPriority, int* greatestPriority ) {
assert(0);
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipDeviceReset ( void ) {
@@ -351,7 +351,7 @@ hipError_t hipDeviceReset ( void ) {
/* FIXME */
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipDeviceSetCacheConfig ( hipFuncCache_t cacheConfig ) {
@@ -359,11 +359,11 @@ hipError_t hipDeviceSetCacheConfig ( hipFuncCache_t cacheConfig ) {
// No way to set cache config yet.
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipDeviceSetLimit ( hipLimit_t limit, size_t value ) {
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
hipError_t hipDeviceSetSharedMemConfig ( hipSharedMemConfig config ) {
@@ -371,7 +371,7 @@ hipError_t hipDeviceSetSharedMemConfig ( hipSharedMemConfig config ) {
// No way to set cache config yet.
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipDeviceSynchronize ( void ) {
@@ -382,11 +382,11 @@ hipError_t hipDeviceSynchronize ( void ) {
amd::HostQueue* queue = hip::getNullStream();
if (!queue) {
return hipErrorOutOfMemory;
HIP_RETURN(hipErrorOutOfMemory);
}
queue->finish();
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipGetDevice ( int* deviceId ) {
@@ -396,24 +396,24 @@ hipError_t hipGetDevice ( int* deviceId ) {
for (unsigned int i = 0; i < g_devices.size(); i++) {
if (g_devices[i] == hip::getCurrentContext()) {
*deviceId = i;
return hipSuccess;
HIP_RETURN(hipSuccess);
}
}
} else {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
hipError_t hipGetDeviceCount ( int* count ) {
HIP_INIT_API(count);
return ihipDeviceGetCount(count);
HIP_RETURN(ihipDeviceGetCount(count));
}
hipError_t hipGetDeviceFlags ( unsigned int* flags ) {
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
hipError_t hipSetDevice ( int device ) {
@@ -438,9 +438,9 @@ hipError_t hipSetDevice ( int device ) {
if (static_cast<unsigned int>(device) < g_devices.size()) {
hip::setCurrentContext(device);
return hipSuccess;
HIP_RETURN(hipSuccess);
}
return hipErrorInvalidDevice;
HIP_RETURN(hipErrorInvalidDevice);
}
hipError_t hipSetDeviceFlags ( unsigned int flags ) {
@@ -453,10 +453,10 @@ hipError_t hipSetDeviceFlags ( unsigned int flags ) {
hipDeviceScheduleMask | hipDeviceMapHost | hipDeviceLmemResizeToMax;
if (flags & (~supportedFlags)) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipSetValidDevices ( int* device_arr, int len ) {
@@ -464,6 +464,6 @@ hipError_t hipSetValidDevices ( int* device_arr, int len ) {
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
projects/clr/hipamd/api/hip/hip_error.cpp
+128 -4
View File
@@ -27,22 +27,146 @@ THE SOFTWARE.
hipError_t hipGetLastError()
{
HIP_INIT_API();
return hipErrorUnknown;
hipError_t err = hip::g_lastError;
hip::g_lastError = hipSuccess;
return err;
}
hipError_t hipPeekAtLastError()
{
HIP_INIT_API();
return hipErrorUnknown;
hipError_t err = hip::g_lastError;
HIP_RETURN(err);
}
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)
{
return "";
return hipGetErrorName(hip_error);
}
projects/clr/hipamd/api/hip/hip_event.cpp
+17 -17
View File
@@ -64,38 +64,38 @@ hipError_t ihipEventQuery(hipEvent_t event) {
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) {
HIP_INIT_API(event, flags);
return ihipEventCreateWithFlags(event, flags);
HIP_RETURN(ihipEventCreateWithFlags(event, flags));
}
hipError_t hipEventCreate(hipEvent_t* event) {
HIP_INIT_API(event);
return ihipEventCreateWithFlags(event, 0);
HIP_RETURN(ihipEventCreateWithFlags(event, 0));
}
hipError_t hipEventDestroy(hipEvent_t event) {
HIP_INIT_API(event);
if (event == nullptr) {
return hipErrorInvalidResourceHandle;
HIP_RETURN(hipErrorInvalidResourceHandle);
}
delete reinterpret_cast<hip::Event*>(event);
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipEventElapsedTime(float *ms, hipEvent_t start, hipEvent_t stop) {
HIP_INIT_API(ms, start, stop);
if (start == nullptr || stop == nullptr) {
return hipErrorInvalidResourceHandle;
HIP_RETURN(hipErrorInvalidResourceHandle);
}
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 ||
eStop->event_ == nullptr) {
return hipErrorInvalidResourceHandle;
HIP_RETURN(hipErrorInvalidResourceHandle);
}
if ((eStart->flags | eStop->flags) & hipEventDisableTiming) {
return hipErrorInvalidResourceHandle;
HIP_RETURN(hipErrorInvalidResourceHandle);
}
if (ihipEventQuery(start) == hipErrorNotReady ||
ihipEventQuery(stop) == hipErrorNotReady) {
return hipErrorNotReady;
HIP_RETURN(hipErrorNotReady);
}
if (ms == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
*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) {
HIP_INIT_API(event, stream);
if (event == nullptr) {
return hipErrorInvalidResourceHandle;
HIP_RETURN(hipErrorInvalidResourceHandle);
}
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();
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipEventSynchronize(hipEvent_t event) {
HIP_INIT_API(event);
if (event == nullptr) {
return hipErrorInvalidResourceHandle;
HIP_RETURN(hipErrorInvalidResourceHandle);
}
hip::Event* e = reinterpret_cast<hip::Event*>(event);
if (e->event_ == nullptr) {
return hipErrorInvalidResourceHandle;
HIP_RETURN(hipErrorInvalidResourceHandle);
}
e->event_->awaitCompletion();
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipEventQuery(hipEvent_t event) {
HIP_INIT_API(event);
return ihipEventQuery(event);
HIP_RETURN(ihipEventQuery(event));
}
projects/clr/hipamd/api/hip/hip_internal.hpp
+6 -1
View File
@@ -39,7 +39,7 @@ THE SOFTWARE.
#define HIP_INIT_API(...) \
amd::Thread* thread = amd::Thread::current(); \
if (!CL_CHECK_THREAD(thread)) { \
return hipErrorOutOfMemory; \
HIP_RETURN(hipErrorOutOfMemory); \
} \
HIP_INIT();
@@ -51,6 +51,7 @@ class accelerator_view;
namespace hip {
extern std::once_flag g_ihipInitialized;
extern thread_local amd::Context* g_context;
extern thread_local hipError_t g_lastError;
extern void init();
@@ -64,5 +65,9 @@ extern std::vector<amd::Context*> g_devices;
extern hipError_t ihipDeviceGetCount(int* count);
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
projects/clr/hipamd/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) {
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) {
HIP_INIT_API(ptr, sizeBytes, flags);
if (ptr == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
*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
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) {
@@ -177,9 +177,9 @@ hipError_t hipFree(void* ptr) {
hip::syncStreams();
hip::getNullStream()->finish();
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) {
@@ -187,7 +187,7 @@ hipError_t hipMemcpy(void* dst, const void* src, size_t sizeBytes, hipMemcpyKind
hip::syncStreams();
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) {
@@ -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();
}
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) {
@@ -212,7 +212,7 @@ hipError_t hipMemset(void* dst, int value, size_t sizeBytes) {
hip::syncStreams();
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) {
@@ -222,12 +222,12 @@ hipError_t hipMemPtrGetInfo(void *ptr, size_t *size) {
amd::Memory* svmMem = getMemoryObject(ptr, offset);
if (svmMem == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
*size = svmMem->getSize();
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipHostFree(void* ptr) {
@@ -235,9 +235,9 @@ hipError_t hipHostFree(void* ptr) {
if (amd::SvmBuffer::malloced(ptr)) {
amd::SvmBuffer::free(*hip::getCurrentContext(), ptr);
return hipSuccess;
HIP_RETURN(hipSuccess);
}
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
hipError_t hipFreeArray(hipArray* array) {
@@ -245,9 +245,9 @@ hipError_t hipFreeArray(hipArray* array) {
if (amd::SvmBuffer::malloced(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) {
@@ -259,13 +259,13 @@ hipError_t hipMemGetAddressRange(hipDeviceptr_t* pbase, size_t* psize, hipDevice
amd::Memory* svmMem = getMemoryObject(ptr, offset);
if (svmMem == nullptr) {
return hipErrorInvalidDevicePointer;
HIP_RETURN(hipErrorInvalidDevicePointer);
}
*pbase = svmMem->getSvmPtr();
*psize = svmMem->getSize();
return hipSuccess;
HIP_RETURN(hipSuccess);
}
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];
amd::Device* device = hip::getCurrentContext()->devices()[0];
if(device == nullptr) {
return hipErrorInvalidDevice;
HIP_RETURN(hipErrorInvalidDevice);
}
if(!device->globalFreeMemory(freeMemory)) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
*free = freeMemory[0] * Ki;
*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,
@@ -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);
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) {
@@ -334,7 +334,7 @@ hipError_t hipMalloc3D(hipPitchedPtr* pitchedDevPtr, hipExtent extent) {
size_t pitch = 0;
if (pitchedDevPtr == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
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;
}
return status;
HIP_RETURN(status);
}
hipError_t hipMemset3D(hipPitchedPtr pitchedDevPtr, int value, hipExtent extent) {
@@ -360,14 +360,14 @@ hipError_t hipMemset3D(hipPitchedPtr pitchedDevPtr, int value, hipExtent extent)
hip::syncStreams();
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) {
HIP_INIT_API(array, pAllocateArray);
if (array[0]->width == 0) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
*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,
&image_format);
return status;
HIP_RETURN(status);
}
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);
if (width == 0) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
*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,
&image_format);
return status;
HIP_RETURN(status);
}
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,
CL_MEM_OBJECT_IMAGE3D, &image_format);
return status;
HIP_RETURN(status);
}
hipError_t hipHostGetFlags(unsigned int* flagsPtr, void* hostPtr) {
@@ -483,19 +483,19 @@ hipError_t hipHostGetFlags(unsigned int* flagsPtr, void* hostPtr) {
if (flagsPtr == nullptr ||
hostPtr == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
size_t offset = 0;
amd::Memory* svmMem = getMemoryObject(hostPtr, offset);
if (svmMem == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
*flagsPtr = svmMem->getMemFlags() >> 16;
return hipSuccess;
HIP_RETURN(hipSuccess);
}
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)) {
mem->release();
return hipErrorMemoryAllocation;
HIP_RETURN(hipErrorMemoryAllocation);
}
amd::MemObjMap::AddMemObj(hostPtr, mem);
return hipSuccess;
HIP_RETURN(hipSuccess);
} else {
return ihipMalloc(&hostPtr, sizeBytes, flags);
HIP_RETURN(ihipMalloc(&hostPtr, sizeBytes, flags));
}
}
@@ -522,7 +522,7 @@ hipError_t hipHostUnregister(void* hostPtr) {
hip::syncStreams();
hip::getNullStream()->finish();
amd::SvmBuffer::free(*hip::getCurrentContext(), hostPtr);
return hipSuccess;
HIP_RETURN(hipSuccess);
} else {
size_t offset = 0;
amd::Memory* mem = getMemoryObject(hostPtr, offset);
@@ -532,16 +532,16 @@ hipError_t hipHostUnregister(void* hostPtr) {
hip::getNullStream()->finish();
amd::MemObjMap::RemoveMemObj(hostPtr);
mem->release();
return hipSuccess;
HIP_RETURN(hipSuccess);
}
}
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
// Deprecated function:
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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();
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) {
@@ -596,7 +596,7 @@ hipError_t hipMemcpyDtoH(void* dst, hipDeviceptr_t src, size_t sizeBytes) {
hip::syncStreams();
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) {
@@ -605,7 +605,7 @@ hipError_t hipMemcpyDtoD(hipDeviceptr_t dst, hipDeviceptr_t src, size_t sizeByte
hip::syncStreams();
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) {
@@ -614,7 +614,7 @@ hipError_t hipMemcpyHtoH(void* dst, void* src, size_t sizeBytes) {
hip::syncStreams();
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,
@@ -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();
}
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();
}
return ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyHostToDevice,
*queue, true);
HIP_RETURN(ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyHostToDevice,
*queue, true));
}
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();
}
return ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyDeviceToDevice,
*queue, true);
HIP_RETURN(ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyDeviceToDevice,
*queue, true));
}
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();
}
return ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyDeviceToHost,
*queue, true);
HIP_RETURN(ihipMemcpy(reinterpret_cast<void*>(dst), (const void*) src, sizeBytes, hipMemcpyDeviceToHost,
*queue, true));
}
hipError_t hipMemcpyParam2D(const hip_Memcpy2D* pCopy) {
@@ -694,7 +694,7 @@ hipError_t hipMemcpyParam2D(const hip_Memcpy2D* pCopy) {
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,
@@ -763,7 +763,7 @@ hipError_t hipMemcpy2D(void* dst, size_t dpitch, const void* src, size_t spitch,
hip::syncStreams();
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();
}
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,
@@ -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);
if (dst->data == nullptr) {
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
hip::syncStreams();
@@ -816,7 +816,7 @@ hipError_t hipMemcpy2DToArray(hipArray* dst, size_t wOffset, size_t hOffset, con
}
if ((wOffset + width > (dpitch)) || width > spitch) {
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
// 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) ||
!dstRect.create(dOrigin, region, dpitch, dst_slice_pitch)) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
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);
memcpy(pDst, pSrc, width);
}
return hipSuccess;
HIP_RETURN(hipSuccess);
} else if ((srcMemory == nullptr) && (dstMemory != nullptr)) {
command = new amd::WriteMemoryCommand(*queue, CL_COMMAND_WRITE_BUFFER_RECT, waitList,
*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) {
return hipErrorOutOfMemory;
HIP_RETURN(hipErrorOutOfMemory);
}
command->enqueue();
command->awaitCompletion();
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) {
return hipErrorOutOfMemory;
HIP_RETURN(hipErrorOutOfMemory);
}
command->enqueue();
command->awaitCompletion();
command->release();
return hipSuccess;
HIP_RETURN(hipSuccess);
}
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) {
return hipErrorOutOfMemory;
HIP_RETURN(hipErrorOutOfMemory);
}
command->enqueue();
command->awaitCompletion();
command->release();
return hipSuccess;
HIP_RETURN(hipSuccess);
}
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) {
return hipErrorOutOfMemory;
HIP_RETURN(hipErrorOutOfMemory);
}
command->enqueue();
command->awaitCompletion();
command->release();
return hipSuccess;
HIP_RETURN(hipSuccess);
}
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) {
return hipErrorOutOfMemory;
HIP_RETURN(hipErrorOutOfMemory);
}
command->enqueue();
command->awaitCompletion();
command->release();
return hipSuccess;
HIP_RETURN(hipSuccess);
}
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) ||
!dstRect.create(dstOrigin, region, dstPitchInbytes, dst_slice_pitch)) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
hipMemcpyKind kind = p->kind;
@@ -1123,7 +1123,7 @@ hipError_t hipMemcpy3D(const struct hipMemcpy3DParms* p) {
if (((srcMemory == nullptr) && (dstMemory == nullptr)) ||
(kind == hipMemcpyHostToHost)) {
memcpy(dstPtr, srcPtr, region[0] * region[1] * region[2]);
return hipSuccess;
HIP_RETURN(hipSuccess);
} else if ((srcMemory == nullptr) && (dstMemory != nullptr)) {
command = new amd::WriteMemoryCommand(*queue, CL_COMMAND_WRITE_BUFFER_RECT, waitList,
*dstMemory->asBuffer(), srcStart, size, srcPtr, srcRect, dstRect);
@@ -1137,14 +1137,14 @@ hipError_t hipMemcpy3D(const struct hipMemcpy3DParms* p) {
}
if (command == nullptr) {
return hipErrorOutOfMemory;
HIP_RETURN(hipErrorOutOfMemory);
}
command->enqueue();
command->awaitCompletion();
command->release();
return hipSuccess;
HIP_RETURN(hipSuccess);
}
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();
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,
@@ -1208,14 +1208,14 @@ hipError_t hipMemset2DAsync(void* dst, size_t pitch, int value,
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) {
HIP_INIT_API(dst, value, sizeBytes);
if (dst == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
hip::syncStreams();
@@ -1232,7 +1232,7 @@ hipError_t hipMemsetD8(hipDeviceptr_t dst, unsigned char value, size_t sizeBytes
&value, sizeof(char), fillOffset, fillSize);
if (command == nullptr) {
return hipErrorOutOfMemory;
HIP_RETURN(hipErrorOutOfMemory);
}
command->enqueue();
@@ -1243,7 +1243,7 @@ hipError_t hipMemsetD8(hipDeviceptr_t dst, unsigned char value, size_t sizeBytes
memset(dst, value, sizeBytes);
}
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipIpcGetMemHandle(hipIpcMemHandle_t* handle, void* devPtr) {
@@ -1251,7 +1251,7 @@ hipError_t hipIpcGetMemHandle(hipIpcMemHandle_t* handle, void* devPtr) {
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
hipError_t hipIpcCloseMemHandle(void* devPtr) {
@@ -1267,7 +1267,7 @@ hipError_t hipIpcCloseMemHandle(void* devPtr) {
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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);
if (!memObj) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
*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) {
@@ -1325,5 +1325,5 @@ hipError_t hipPointerGetAttributes(hipPointerAttribute_t* attributes, const void
attributes->allocationFlags = 0;
}
return hipSuccess;
HIP_RETURN(hipSuccess);
}
projects/clr/hipamd/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);
if (!fname) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
std::ifstream file{fname};
if (!file.is_open()) {
return hipErrorFileNotFound;
HIP_RETURN(hipErrorFileNotFound);
}
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);
if (hmod == nullptr) {
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
amd::Program* program = as_amd(reinterpret_cast<cl_program>(hmod));
program->release();
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipModuleLoadData(hipModule_t *module, const void *image)
{
HIP_INIT_API(module, image);
return ihipModuleLoadData(module, image);
HIP_RETURN(ihipModuleLoadData(module, 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);
if (!symbol) {
return hipErrorNotFound;
HIP_RETURN(hipErrorNotFound);
}
amd::Kernel* kernel = new amd::Kernel(*program, *symbol, name);
if (!kernel) {
return hipErrorOutOfMemory;
HIP_RETURN(hipErrorOutOfMemory);
}
*hfunc = reinterpret_cast<hipFunction_t>(as_cl(kernel));
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipFuncGetAttributes(hipFuncAttributes* attr, const void* 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,
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,
sharedMemBytes, hStream, kernelParams, extra, nullptr, nullptr);
sharedMemBytes, hStream, kernelParams, extra, nullptr, nullptr));
}
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 stopEvent)
{
return ihipModuleLaunchKernel(f, gridDimX, gridDimY, gridDimZ, blockDimX, blockDimY, blockDimZ,
sharedMemBytes, hStream, kernelParams, extra, startEvent, stopEvent);
HIP_RETURN(ihipModuleLaunchKernel(f, gridDimX, gridDimY, gridDimZ, blockDimX, blockDimY, blockDimZ,
sharedMemBytes, hStream, kernelParams, extra, startEvent, stopEvent));
}
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 stopEvent)
{
return ihipModuleLaunchKernel(f, gridDimX, gridDimY, gridDimZ, blockDimX, blockDimY, blockDimZ,
sharedMemBytes, hStream, kernelParams, extra, startEvent, stopEvent);
HIP_RETURN(ihipModuleLaunchKernel(f, gridDimX, gridDimY, gridDimZ, blockDimX, blockDimY, blockDimZ,
sharedMemBytes, hStream, kernelParams, extra, startEvent, stopEvent));
}
projects/clr/hipamd/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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
hipError_t hipDeviceDisablePeerAccess(int peerDeviceId) {
@@ -63,7 +63,7 @@ hipError_t hipDeviceDisablePeerAccess(int peerDeviceId) {
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
hipError_t hipDeviceEnablePeerAccess(int peerDeviceId, unsigned int flags) {
@@ -71,7 +71,7 @@ hipError_t hipDeviceEnablePeerAccess(int peerDeviceId, unsigned int flags) {
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
hipError_t hipCtxDisablePeerAccess(hipCtx_t peerCtx) {
@@ -105,5 +105,5 @@ hipError_t hipCtxDisablePeerAccess(hipCtx_t peerCtx) {
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
projects/clr/hipamd/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);
return hipSuccess;
HIP_RETURN(hipSuccess);
}
extern "C" hipError_t hipSetupArgument(
@@ -264,7 +264,7 @@ extern "C" hipError_t hipSetupArgument(
PlatformState::instance().setupArgument(arg, size, offset);
return hipSuccess;
HIP_RETURN(hipSuccess);
}
extern "C" hipError_t hipLaunchByPtr(const void *hostFunction)
@@ -272,8 +272,9 @@ extern "C" hipError_t hipLaunchByPtr(const void *hostFunction)
HIP_INIT_API(hostFunction);
hipFunction_t func = PlatformState::instance().getFunc(hostFunction);
if (func == nullptr)
return hipErrorUnknown;
if (func == nullptr) {
HIP_RETURN(hipErrorUnknown);
}
ihipExec_t exec;
PlatformState::instance().popExec(exec);
@@ -284,10 +285,10 @@ extern "C" hipError_t hipLaunchByPtr(const void *hostFunction)
HIP_LAUNCH_PARAM_END
};
return hipModuleLaunchKernel(func,
HIP_RETURN(hipModuleLaunchKernel(func,
exec.gridDim_.x, exec.gridDim_.y, exec.gridDim_.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)
@@ -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 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 {
std::uint32_t u = ((a << 13) + 0x70000000U) & 0x8fffe000U;
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;
}
// 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
static inline std::uint32_t __convert_float_to_half(float a) noexcept {
std::uint32_t u = f32_as_u32(a);
projects/clr/hipamd/api/hip/hip_profile.cpp
+2 -2
View File
@@ -29,7 +29,7 @@ hipError_t hipProfilerStart() {
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
@@ -38,5 +38,5 @@ hipError_t hipProfilerStop() {
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
projects/clr/hipamd/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) {
HIP_INIT_API(stream, flags);
return ihipStreamCreateWithFlags(stream, flags);
HIP_RETURN(ihipStreamCreateWithFlags(stream, flags));
}
hipError_t hipStreamCreate(hipStream_t *stream) {
HIP_INIT_API(stream);
return ihipStreamCreateWithFlags(stream, hipStreamDefault);
HIP_RETURN(ihipStreamCreateWithFlags(stream, hipStreamDefault));
}
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) {
*flags = (it == streamSet.end()) ? hipStreamNonBlocking : hipStreamDefault;
} else {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipStreamSynchronize(hipStream_t stream) {
@@ -108,19 +108,19 @@ hipError_t hipStreamSynchronize(hipStream_t stream) {
}
if (hostQueue == nullptr) {
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
hostQueue->finish();
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipStreamDestroy(hipStream_t stream) {
HIP_INIT_API(stream);
if (stream == nullptr) {
return hipErrorInvalidResourceHandle;
HIP_RETURN(hipErrorInvalidResourceHandle);
}
amd::ScopedLock lock(streamSetLock);
@@ -129,14 +129,14 @@ hipError_t hipStreamDestroy(hipStream_t stream) {
hostQueue->release();
streamSet.erase(hostQueue);
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipStreamWaitEvent(hipStream_t stream, hipEvent_t event, unsigned int flags) {
HIP_INIT_API(stream, event, flags);
if (stream == nullptr || event == nullptr) {
return hipErrorInvalidResourceHandle;
HIP_RETURN(hipErrorInvalidResourceHandle);
}
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;
cl_int err = amd::clSetEventWaitList(eventWaitList, *hostQueue, 1, &clEvent);
if (err != CL_SUCCESS) {
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
amd::Command* command = new amd::Marker(*hostQueue, true, eventWaitList);
if (command == NULL) {
return hipErrorOutOfMemory;
HIP_RETURN(hipErrorOutOfMemory);
}
command->enqueue();
command->release();
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipStreamQuery(hipStream_t stream) {
@@ -164,7 +164,7 @@ hipError_t hipStreamQuery(hipStream_t stream) {
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
projects/clr/hipamd/api/hip/hip_surface.cpp
+2 -2
View File
@@ -36,7 +36,7 @@ hipError_t hipCreateSurfaceObject(hipSurfaceObject_t* pSurfObject,
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
@@ -45,5 +45,5 @@ hipError_t hipDestroySurfaceObject(hipSurfaceObject_t surfaceObject) {
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
projects/clr/hipamd/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];
if (!device->info().imageSupport_) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
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.pitchInBytes, 0);
break;
default: return hipErrorInvalidValue;
default: HIP_RETURN(hipErrorInvalidValue);
}
*pTexObject = reinterpret_cast<hipTextureObject_t>(as_cl(image));
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
hipError_t hipDestroyTextureObject(hipTextureObject_t textureObject) {
@@ -215,7 +215,7 @@ hipError_t hipDestroyTextureObject(hipTextureObject_t textureObject) {
as_amd(reinterpret_cast<cl_mem>(textureObject))->release();
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipGetTextureObjectResourceDesc(hipResourceDesc* pResDesc,
@@ -224,7 +224,7 @@ hipError_t hipGetTextureObjectResourceDesc(hipResourceDesc* pResDesc,
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
hipError_t hipGetTextureObjectResourceViewDesc(hipResourceViewDesc* pResViewDesc,
@@ -233,7 +233,7 @@ hipError_t hipGetTextureObjectResourceViewDesc(hipResourceViewDesc* pResViewDesc
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
hipError_t hipGetTextureObjectTextureDesc(hipTextureDesc* pTexDesc,
@@ -242,7 +242,7 @@ hipError_t hipGetTextureObjectTextureDesc(hipTextureDesc* pTexDesc,
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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);
if (desc == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
cl_image_format image_format;
getChannelOrderAndType(*desc, hipReadModeElementType,
&image_format.image_channel_order, &image_format.image_channel_data_type);
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,
@@ -297,7 +297,7 @@ hipError_t hipBindTexture2D(size_t* offset, textureReference* tex, const void* d
size_t 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,
@@ -306,7 +306,7 @@ hipError_t hipBindTextureToArray(textureReference* tex, hipArray_const_t array,
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
hipError_t ihipBindTextureToArrayImpl(int dim, enum hipTextureReadMode readMode,
@@ -325,7 +325,7 @@ hipError_t hipBindTextureToMipmappedArray(textureReference* tex,
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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();
return hipSuccess;
HIP_RETURN(hipSuccess);
}
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
hipError_t hipGetTextureAlignmentOffset(size_t* offset, const textureReference* tex) {
@@ -349,7 +349,7 @@ hipError_t hipGetTextureAlignmentOffset(size_t* offset, const textureReference*
assert(0 && "Unimplemented");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
hipError_t hipTexRefSetFormat(textureReference* tex, hipArray_Format fmt, int NumPackedComponents) {
HIP_INIT_API(tex, fmt, NumPackedComponents);
if (tex == nullptr) {
return hipErrorInvalidImage;
HIP_RETURN(hipErrorInvalidImage);
}
tex->format = fmt;
tex->numChannels = NumPackedComponents;
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipTexRefSetFlags(textureReference* tex, unsigned int flags) {
HIP_INIT_API(tex, flags);
if (tex == nullptr) {
return hipErrorInvalidImage;
HIP_RETURN(hipErrorInvalidImage);
}
tex->normalized = flags;
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipTexRefSetFilterMode(textureReference* tex, hipTextureFilterMode fm) {
HIP_INIT_API(tex, fm);
if (tex == nullptr) {
return hipErrorInvalidImage;
HIP_RETURN(hipErrorInvalidImage);
}
tex->filterMode = fm;
return hipSuccess;
HIP_RETURN(hipSuccess);
}
hipError_t hipTexRefSetAddressMode(textureReference* tex, int dim, hipTextureAddressMode am) {
HIP_INIT_API(tex, dim, am);
if (tex == nullptr) {
return hipErrorInvalidImage;
HIP_RETURN(hipErrorInvalidImage);
}
tex->addressMode[dim] = am;
return hipSuccess;
HIP_RETURN(hipSuccess);
}
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");
return hipErrorUnknown;
HIP_RETURN(hipErrorUnknown);
}
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);
if (tex == nullptr) {
return hipErrorInvalidImage;
HIP_RETURN(hipErrorInvalidImage);
}
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);
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,
@@ -438,9 +438,9 @@ hipError_t hipTexRefSetAddress2D(textureReference* tex, const HIP_ARRAY_DESCRIPT
HIP_INIT_API(tex, desc, devPtr, pitch);
if (desc == nullptr) {
return hipErrorInvalidValue;
HIP_RETURN(hipErrorInvalidValue);
}
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));
}