From 11edc27b8799c694d463fd6f9871e660d3de2c44 Mon Sep 17 00:00:00 2001 From: sunway513 Date: Thu, 28 Jan 2016 10:59:21 -0600 Subject: [PATCH] Fix @file and @brief tag on header files [ROCm/hip commit: 71a841d764e4b7de0e648995a6a9998a7148218e] --- .../hip/include/hcc_detail/hip_runtime_api.h | 170 +++++++++--------- projects/hip/include/hcc_detail/hip_texture.h | 42 ++--- .../hip/include/hcc_detail/hip_vector_types.h | 6 + .../hip/include/hcc_detail/host_defines.h | 24 ++- 4 files changed, 129 insertions(+), 113 deletions(-) diff --git a/projects/hip/include/hcc_detail/hip_runtime_api.h b/projects/hip/include/hcc_detail/hip_runtime_api.h index 63238ff354..e9dbd4e626 100644 --- a/projects/hip/include/hcc_detail/hip_runtime_api.h +++ b/projects/hip/include/hcc_detail/hip_runtime_api.h @@ -21,20 +21,24 @@ THE SOFTWARE. */ #pragma once +/** + * @file hcc_detail/hip_runtime_api.h + * @brief Contains C function APIs for HIP runtime. + */ #include #include #include -#if defined (__HCC__) && (__hcc_workweek__ < 1602) +#if defined (__HCC__) && (__hcc_workweek__ < 1602) #error("This version of HIP requires a newer version of HCC."); #endif // hip_api_hcc.h // Contains C function APIs for HIP runtime. -// This file does not use any HCC builtins or special language extensions (-hc mode) ; those functions in hip_hcc.h. +// This file does not use any HCC builtin or special language extensions (-hc mode) ; those functions in hip_hcc.h. // Structure definitions: @@ -43,18 +47,18 @@ extern "C" { #endif /** - * @addtogroup GlobalDefs More + * @addtogroup GlobalDefs More * @{ */ //! Flags that can be used with hipStreamCreateWithFlags #define hipStreamDefault 0x00 ///< Default stream creation flags. These are used with hipStreamCreate(). -#define hipStreamNonBlocking 0x01 ///< Stream does not implicitly synchronize with null stream +#define hipStreamNonBlocking 0x01 ///< Stream does not implicitly synchronize with null stream //! Flags that can be used with hipEventCreateWithFlags: #define hipEventDefault 0x0 ///< Default flags #define hipEventBlockingSync 0x1 ///< Waiting will yield CPU. Power-friendly and usage-friendly but may increase latency. -#define hipEventDisableTiming 0x2 ///< Disable event's capability to record timing information. May improve performance. +#define hipEventDisableTiming 0x2 ///< Disable event's capability to record timing information. May improve performance. #define hipEventInterprocess 0x4 ///< Event can support IPC. @warning - not supported in HIP. @@ -118,11 +122,11 @@ enum hipMemcpyKind { typedef struct ihipStream_t * hipStream_t; -/* +/* * Opaque structure allows the true event (pointed at by the handle) to remain "live" even if the surrounding hipEvent_t goes out-of-scope. * This is handy for cases where the hipEvent_t goes out-of-scope but the true event is being written by some async queue or device */ typedef struct hipEvent_t { - struct ihipEvent_t *_handle; + struct ihipEvent_t *_handle; } hipEvent_t; @@ -158,8 +162,8 @@ extern "C" { * @{ */ -/** - * @brief Blocks until the default device has completed all preceding requested tasks. +/** + * @brief Blocks until the default device has completed all preceding requested tasks. * * This function waits for all streams on the default device to complete execution, and then returns. * @@ -169,10 +173,10 @@ hipError_t hipDeviceSynchronize(void); -/** +/** * @brief Destroy all resources and reset all state on the default device in the current process. * - * Explicity destroy all memory allocations, events, and queues associated with the default device in the current process. + * Explicity destroy all memory allocations, events, and queues associated with the default device in the current process. * * This function will reset the device immmediately, and then return after all resources have been freed. * The caller must ensure that the device is not being accessed by any other host threads from the active process when this function is called. @@ -185,23 +189,23 @@ hipError_t hipDeviceReset(void) ; /** * @brief Set default device to be used for subsequent hip API calls from this thread. * - * @param[in] device Valid device in range 0...hipGetDeviceCount(). + * @param[in] device Valid device in range 0...hipGetDeviceCount(). * * Sets @p device as the default device for the calling host thread. Valid device id's are 0... (hipGetDeviceCount()-1). * * Many HIP APIs implicitly use the "default device" : * - * - Any device memory subsequently allocated from this host thread (using hipMalloc) will be allocated on device. + * - Any device memory subsequently allocated from this host thread (using hipMalloc) will be allocated on device. * - Any streams or events created from this host thread will be associated with device. - * - Any kernels launched from this host thread (using hipLaunchKernel) will be executed on device (unless a specific stream is specified, + * - Any kernels launched from this host thread (using hipLaunchKernel) will be executed on device (unless a specific stream is specified, * in which case the device associated with that stream will be used). * - * This function may be called from any host thread. Multiple host threads may use the same device. - * This function does no synchronization with the previous or new device, and has very little runtime overhead. + * This function may be called from any host thread. Multiple host threads may use the same device. + * This function does no synchronization with the previous or new device, and has very little runtime overhead. * Applications can use hipSetDevice to quickly switch the default device before making a HIP runtime call which uses the default device. * - * The default device is stored in thread-local-storage for each thread. - * Thread-pool implementations may inherit the default device of the previous thread. A good practice is to always call hipSetDevice + * The default device is stored in thread-local-storage for each thread. + * Thread-pool implementations may inherit the default device of the previous thread. A good practice is to always call hipSetDevice * at the start of HIP coding sequency to establish a known standard device. * * @see hipGetDevice, hipGetDeviceCount @@ -216,7 +220,7 @@ hipError_t hipSetDevice(int device); * * HIP maintains an default device for each thread using thread-local-storage. * This device is used implicitly for HIP runtime APIs called by this thread. - * hipGetDevice returns in * @p device the default device for the calling host thread. + * hipGetDevice returns in * @p device the default device for the calling host thread. * * @see hipSetDevice, hipGetDevicesizeBytes */ @@ -224,17 +228,17 @@ hipError_t hipGetDevice(int *device); /** - * @brief Return number of compute-capable devices. - * @param [output] count Returns number of compute-capable devices. + * @brief Return number of compute-capable devices. + * @param [output] count Returns number of compute-capable devices. * - * Returns in @p *count the number of devices that have ability to run compute commands. If there are no such devices, then @ref hipGetDeviceCount will return #hipErrorNoDevice. + * Returns in @p *count the number of devices that have ability to run compute commands. If there are no such devices, then @ref hipGetDeviceCount will return #hipErrorNoDevice. * If 1 or more devices can be found, then hipGetDeviceCount returns #hipSuccess. */ hipError_t hipGetDeviceCount(int *count); /** - * @brief Returns device properties. + * @brief Returns device properties. * * @param [out] prop written with device properties * @param [in] device which device to query for information @@ -249,7 +253,7 @@ hipError_t hipDeviceGetProperties(hipDeviceProp_t* prop, int device); /** * @brief Set L1/Shared cache partition. - * + * * Note: AMD devices and recent NVIDIA GPUS do not support reconfigurable cache. This hint is ignored on those architectures. * */ @@ -258,7 +262,7 @@ hipError_t hipDeviceSetCacheConfig ( hipFuncCache cacheConfig ); /** * @brief Set Cache configuration for a specific function - * + * * Note: AMD devices and recent NVIDIA GPUS do not support reconfigurable cache. This hint is ignored on those architectures. * */ @@ -267,31 +271,31 @@ hipError_t hipDeviceGetCacheConfig ( hipFuncCache *cacheConfig ); /** * @brief Set Cache configuration for a specific function - * + * * Note: AMD devices and recent NVIDIA GPUS do not support reconfigurable cache. This hint is ignored on those architectures. * */ -hipError_t hipFuncSetCacheConfig ( hipFuncCache config ); +hipError_t hipFuncSetCacheConfig ( hipFuncCache config ); -//--- +//--- //Shared bank config functions: /** * @brief Get Shared memory bank configuration. - * + * * Note: AMD devices and recent NVIDIA GPUS do not support shared cache banking, and the hint is ignored on those architectures. * */ -hipError_t hipDeviceGetSharedMemConfig ( hipSharedMemConfig * pConfig ); +hipError_t hipDeviceGetSharedMemConfig ( hipSharedMemConfig * pConfig ); /** * @brief Set Shared memory bank configuration. - * + * * Note: AMD devices and recent NVIDIA GPUS do not support shared cache banking, and the hint is ignored on those architectures. * */ -hipError_t hipDeviceSetSharedMemConfig ( hipSharedMemConfig config ); +hipError_t hipDeviceSetSharedMemConfig ( hipSharedMemConfig config ); // end doxygen Device @@ -313,21 +317,21 @@ hipError_t hipDeviceSetSharedMemConfig ( hipSharedMemConfig config ); * Returns the last error that has been returned by any of the runtime calls in the same host thread, * and then resets the saved error to #hipSuccess. * - */ + */ hipError_t hipGetLastError(void); /** - * @brief Return last error returned by any HIP runtime API call. + * @brief Return last error returned by any HIP runtime API call. * * @return #hipSuccess * - * Returns the last error that has been returned by any of the runtime calls in the same host thread. + * Returns the last error that has been returned by any of the runtime calls in the same host thread. * Unlike hipGetLastError, this function does not reset the saved error code. * * * - */ + */ hipError_t hipPeekAtLastError(void); @@ -357,7 +361,7 @@ const char *hipGetErrorString(hipError_t hip_error); // end doxygen Error /** - * @} + * @} */ @@ -373,7 +377,7 @@ const char *hipGetErrorString(hipError_t hip_error); * - cudaStreamGetPriority */ -/** +/** * @brief Create an asynchronous stream. * * @param[in, out] stream Pointer to new stream @@ -389,7 +393,7 @@ hipError_t hipStreamCreateWithFlags(hipStream_t *stream, unsigned int flags); -/** +/** * @brief Create an asynchronous stream. * * @param[in, out] stream Valid pointer to hipStream_t. This function writes the memory with the newly created stream. @@ -398,13 +402,13 @@ hipError_t hipStreamCreateWithFlags(hipStream_t *stream, unsigned int flags); * Create a new asynchronous stream. * */ -static inline hipError_t hipStreamCreate(hipStream_t *stream) +static inline hipError_t hipStreamCreate(hipStream_t *stream) { return hipStreamCreateWithFlags(stream, hipStreamDefault); } -/** +/** * @brief Make the specified compute stream wait for an event * * @param[in] stream stream to make wait. @@ -413,7 +417,7 @@ static inline hipError_t hipStreamCreate(hipStream_t *stream) * * @return #hipSuccess, #hipErrorInvalidResourceHandle * - * This function inserts a wait operation into the specified stream. + * This function inserts a wait operation into the specified stream. * All future work submitted to @p stream will wait until @p event reports completion before beginning execution. * This function is host-asynchronous and the function may return before the wait has completed. * @@ -422,10 +426,10 @@ static inline hipError_t hipStreamCreate(hipStream_t *stream) hipError_t hipStreamWaitEvent(hipStream_t stream, hipEvent_t event, unsigned int flags); -/** +/** * @brief Wait for all commands in stream to complete. * - * If the null stream is specified, this command blocks until all + * If the null stream is specified, this command blocks until all * * This command honors the hipDeviceLaunchBlocking flag, which controls whether the wait is active or blocking. * @@ -436,17 +440,17 @@ hipError_t hipStreamWaitEvent(hipStream_t stream, hipEvent_t event, unsigned int hipError_t hipStreamSynchronize(hipStream_t stream); -/** +/** * @brief Destroys the specified stream. * * @param[in, out] stream Valid pointer to hipStream_t. This function writes the memory with the newly created stream. * @return #hipSuccess * - * Destroys the specified stream. + * Destroys the specified stream. * * If commands are still executing on the specified stream, some may complete execution before the queue is deleted. * - * The queue may be destroyed while some commands are still inflight, or may wait for all commands queued to the stream + * The queue may be destroyed while some commands are still inflight, or may wait for all commands queued to the stream * before destroying it. */ hipError_t hipStreamDestroy(hipStream_t stream); @@ -455,11 +459,11 @@ hipError_t hipStreamDestroy(hipStream_t stream); /** * @brief Return flags associated with this stream. * - * @param[in] stream + * @param[in] stream * @param[in,out] flags * @return #hipSuccess, #hipErrorInvalidValue, #hipErrorInvalidResourceHandle * - * Return flags associated with this stream in *@p flags. + * Return flags associated with this stream in *@p flags. * * @see hipStreamCreateWithFlags * @@ -470,7 +474,7 @@ hipError_t hipStreamGetFlags(hipStream_t stream, unsigned int *flags); // end doxygen Stream /** - * @} + * @} */ @@ -498,7 +502,7 @@ hipError_t hipEventCreateWithFlags(hipEvent_t* event, unsigned flags); /** * Create an event - * + * * @param[in] event Creates an event * */ @@ -515,22 +519,22 @@ static inline hipError_t hipEventCreate(hipEvent_t* event) * @param[in] stream stream in which to record event. * @returns #hipSuccess, #hipErrorInvalidResourceHandle * - * hipEventQuery or hipEventSynchronize must be used to determine when the event + * hipEventQuery or hipEventSynchronize must be used to determine when the event * transitions from "recording" (after eventRecord is called) to "recorded" * (when timestamps are set, if requested). * * Events which are recorded in a non-NULL stream will transition to - * from recording to "recorded" state when they reach the head of - * the specified stream, after all previous - * commands in that stream have completed executing. + * from recording to "recorded" state when they reach the head of + * the specified stream, after all previous + * commands in that stream have completed executing. * * If hipEventRecord has been previously called aon event, then this call will overwrite any existing state in event. * - * If this function is called on a an event that is currently being recorded, results are undefined - either + * If this function is called on a an event that is currently being recorded, results are undefined - either * outstanding recording may save state into the event, and the order is not guaranteed. This shoul be avoided. * * @see hipEventElapsedTime - * + * */ hipError_t hipEventRecord(hipEvent_t event, hipStream_t stream = NULL); @@ -540,11 +544,11 @@ hipError_t hipEventRecord(hipEvent_t event, hipStream_t stream = NULL); * @brief Destroy the specified event. * * @param[in] event Event to destroy. - * @return : #hipSuccess, + * @return : #hipSuccess, * * Releases memory associated with the event. If the event is recording but has not completed recording when hipEventDestroy is called, * the function will return immediately and the completion_future resources will be released later, when the hipDevice is synchronized. - * + * */ hipError_t hipEventDestroy(hipEvent_t event); @@ -559,7 +563,7 @@ hipError_t hipEventDestroy(hipEvent_t event); * TODO-hcc - This function needs to support hipEventBlockingSync parameter. * * @param[in] event Event on which to wait. - * @return #hipSuccess, #hipErrorInvalidResourceHandle, + * @return #hipSuccess, #hipErrorInvalidResourceHandle, * */ hipError_t hipEventSynchronize(hipEvent_t event); @@ -573,15 +577,15 @@ hipError_t hipEventSynchronize(hipEvent_t event); * @param[in] stop : Stop event. * @return : #hipSuccess, #hipErrorInvalidResourceHandle, #hipErrorNotReady, * - * Computes the elapsed time between two events. Time is computed in ms, with - * a resolution of approximately 1 us. + * Computes the elapsed time between two events. Time is computed in ms, with + * a resolution of approximately 1 us. * * Events which are recorded in a NULL stream will block until all commands * on all other streams complete execution, and then record the timestamp. - * + * * Events which are recorded in a non-NULL stream will record their timestamp - * when they reach the head of the specified stream, after all previous - * commands in that stream have completed executing. Thus the time that + * when they reach the head of the specified stream, after all previous + * commands in that stream have completed executing. Thus the time that * the event recorded may be significantly after the host calls hipEventRecord. * * If hipEventRecord has not been called on either event, then #hipErrorInvalidResourceHandle is returned. @@ -600,8 +604,8 @@ hipError_t hipEventElapsedTime(float *ms, hipEvent_t start, hipEvent_t stop); * * Query the status of the specified event. This function will return #hipErrorNotReady if all commands * in the appropriate stream (specified to hipEventRecord) have completed. If that work has not completed, - * or if hipEventRecord was not called on the event, then cudaSuccess is returned. - * + * or if hipEventRecord was not called on the event, then cudaSuccess is returned. + * * */ hipError_t hipEventQuery(hipEvent_t event) ; @@ -609,7 +613,7 @@ hipError_t hipEventQuery(hipEvent_t event) ; // end doxygen Events /** - * @} + * @} */ @@ -642,7 +646,7 @@ hipError_t hipMalloc(void** ptr, size_t size) ; /** * Allocate pinned host memory - * + * * @param[in] ptr Pointer to the allocated host pinned memory * @param[out] size Requested memory size * @return Error code @@ -680,9 +684,9 @@ hipError_t hipFreeHost(void* ptr); * Copy data from src to dst. It supports memory from host to device, * device to host, device to device and host to host * The src and dst must not overlap. - * If the + * If the * - * This function is host-synchronous for most inputs. + * This function is host-synchronous for most inputs. * It uses the default NULL stream and will synchronize with other blocking streams on the same device. * * @param[ being copy to @@ -731,7 +735,7 @@ hipError_t hipMemGetInfo (size_t * free, size_t * total) ; // doxygen end Memory /** - * @} + * @} */ @@ -769,7 +773,7 @@ hipError_t hipMemcpyPeer ( void* dst, int dstDevice, const void* src, int srcD hipError_t hipMemcpyPeerAsync ( void* dst, int dstDevice, const void* src, int srcDevice, size_t sizeBytes, hipStream_t stream=0 ); // doxygen end PeerToPeer /** - * @} + * @} */ @@ -781,12 +785,12 @@ hipError_t hipMemcpyPeerAsync ( void* dst, int dstDevice, const void* src, int * */ -/** +/** * @brief Returns the approximate HIP driver versin. * - * @warning The HIP feature set does not correpond to an exact CUDA SDK driver revision. + * @warning The HIP feature set does not correpond to an exact CUDA SDK driver revision. * This function always set *driverVersion to 4 as an approximation though HIP supports - * some features which were introduced in later CUDA SDK revisions. + * some features which were introduced in later CUDA SDK revisions. * HIP apps code should not rely on the driver revision number here and should * use arch feature flags to test device capabiliies or conditional compilation. * @@ -797,7 +801,7 @@ hipError_t hipDriverGetVersion(int *driverVersion) ; // doxygen end Version Management /** - * @} + * @} */ @@ -806,9 +810,9 @@ hipError_t hipDriverGetVersion(int *driverVersion) ; *------------------------------------------------------------------------------------------------- * @defgroup Profiler Control * @{ - * * - * The cudaProfilerInitialize API format for "configFile" is not supported. + * + * The cudaProfilerInitialize API format for "configFile" is not supported. * * On AMD platforms, hipProfilerStart and hipProfilerStop require installation of AMD's GPU * perf counter API and defining GPU_PERF @@ -816,7 +820,7 @@ hipError_t hipDriverGetVersion(int *driverVersion) ; /** - * @} + * @} */ @@ -832,7 +836,7 @@ hipError_t hipDriverGetVersion(int *driverVersion) ; *------------------------------------------------------------------------------------------------- * @defgroup HCC_Specific HCC-Specific Accessors * @{ - * + * * The following calls are only supported when compiler HIP with HCC. * To produce portable code, use of these calls must be guarded #ifdef checks: * @code @@ -844,7 +848,7 @@ hipError_t hipDriverGetVersion(int *driverVersion) ; * */ -#ifdef __HCC__ +#ifdef __HCC__ #include /** * @brief Return hc::acclerator associated with the specified deviceId @@ -860,7 +864,7 @@ hipError_t hipHccGetAcceleratorView(hipStream_t stream, hc::accelerator_view **a // end-group HCC_Specific /** - * @} + * @} */ diff --git a/projects/hip/include/hcc_detail/hip_texture.h b/projects/hip/include/hcc_detail/hip_texture.h index 0ba13faeb4..b71a54ed14 100644 --- a/projects/hip/include/hcc_detail/hip_texture.h +++ b/projects/hip/include/hcc_detail/hip_texture.h @@ -21,11 +21,11 @@ THE SOFTWARE. */ #pragma once /** - * @file hip_kalmar_texture.h + * @file hcc_detail/hip_texture.h * @brief HIP C++ Texture API for hcc compiler */ -#include +#include #include @@ -38,13 +38,13 @@ typedef struct hipChannelFormatDesc { int _dummy; } hipChannelFormatDesc; -typedef enum hipTextureReadMode +typedef enum hipTextureReadMode { hipReadModeElementType, ///< Read texture as specified element type //! @warning cudaReadModeNormalizedFloat is not supported. } hipTextureReadMode; -typedef enum hipTextureFilterMode +typedef enum hipTextureFilterMode { hipFilterModePoint, ///< Point filter mode. //! @warning cudaFilterModeLinear is not supported. @@ -86,26 +86,26 @@ struct texture : public textureReference { * @defgroup Texture Texture Reference Management * @{ * - * + * * @warning The HIP texture API implements a small subset of full texture API. Known limitations include: * - Only point sampling is supported. * - Only C++ APIs are provided. - * - Many APIs and modes are not implemented. + * - Many APIs and modes are not implemented. * * The HIP texture support is intended to allow use of texture cache on hardware where this is beneficial. * * The following CUDA APIs are not currently supported: - * - cudaBindTexture2D + * - cudaBindTexture2D * - cudaBindTextureToArray - * - cudaBindTextureToMipmappedArray - * - cudaGetChannelDesc + * - cudaBindTextureToMipmappedArray + * - cudaGetChannelDesc * - cudaGetTextureReference * */ // C API: #if 0 -hipChannelFormatDesc hipBindTexture(size_t *offset, struct textureReference *tex, const void *devPtr, const struct hipChannelFormatDesc *desc, size_t size=UINT_MAX) +hipChannelFormatDesc hipBindTexture(size_t *offset, struct textureReference *tex, const void *devPtr, const struct hipChannelFormatDesc *desc, size_t size=UINT_MAX) { tex->_dataPtr = devPtr; } @@ -116,7 +116,7 @@ hipChannelFormatDesc hipBindTexture(size_t *offset, struct textureReference *te **/ // TODO template -hipChannelFormatDesc hipCreateChannelDesc() +hipChannelFormatDesc hipCreateChannelDesc() { hipChannelFormatDesc desc; return desc; @@ -127,11 +127,11 @@ hipChannelFormatDesc hipCreateChannelDesc() **/ // TODO-doc template -hipError_t hipBindTexture(size_t *offset, - struct texture &tex, - const void *devPtr, - const struct hipChannelFormatDesc *desc, - size_t size=UINT_MAX) +hipError_t hipBindTexture(size_t *offset, + struct texture &tex, + const void *devPtr, + const struct hipChannelFormatDesc *desc, + size_t size=UINT_MAX) { tex._dataPtr = static_cast(devPtr); @@ -144,10 +144,10 @@ hipError_t hipBindTexture(size_t *offset, **/ // TODO-doc template -hipError_t hipBindTexture(size_t *offset, - struct texture &tex, - const void *devPtr, - size_t size=UINT_MAX) +hipError_t hipBindTexture(size_t *offset, + struct texture &tex, + const void *devPtr, + size_t size=UINT_MAX) { return hipBindTexture(offset, tex, devPtr, &tex.channelDesc, size); } @@ -169,7 +169,7 @@ hipError_t hipUnbindTexture(struct texture *tex) // doxygen end Texture /** - * @} + * @} */ diff --git a/projects/hip/include/hcc_detail/hip_vector_types.h b/projects/hip/include/hcc_detail/hip_vector_types.h index e448b56131..29d2d53a8d 100644 --- a/projects/hip/include/hcc_detail/hip_vector_types.h +++ b/projects/hip/include/hcc_detail/hip_vector_types.h @@ -19,6 +19,12 @@ LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE SOFTWARE. */ + +/** + * @file hcc_detail/hip_vector_types.h + * @brief Defines the different newt vector types for HIP runtime. + */ + #if defined (__HCC__) && (__hcc_workweek__ < 16032) #error("This version of HIP requires a newer version of HCC."); #endif diff --git a/projects/hip/include/hcc_detail/host_defines.h b/projects/hip/include/hcc_detail/host_defines.h index a400b5fe83..6e12e26e0b 100644 --- a/projects/hip/include/hcc_detail/host_defines.h +++ b/projects/hip/include/hcc_detail/host_defines.h @@ -19,6 +19,12 @@ LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE SOFTWARE. */ + +/** + * @file hcc_detail/host_defines.h + * @brief TODO-doc + */ + #ifdef __HCC__ /** * Function and kernel markers @@ -29,7 +35,7 @@ THE SOFTWARE. #ifndef DISABLE_GRID_LAUNCH #define __global__ __attribute__((hc_grid_launch)) #else -#define __global__ +#define __global__ #endif #define __noinline__ __attribute__((noinline)) @@ -44,20 +50,20 @@ THE SOFTWARE. #define __shared__ tile_static #define __constant__ __attribute__((address_space(2))) -#else +#else // Non-HCC compiler /** * Function and kernel markers */ -#define __host__ -#define __device__ +#define __host__ +#define __device__ -#define __global__ +#define __global__ -#define __noinline__ -#define __forceinline__ +#define __noinline__ +#define __forceinline__ -#define __shared__ -#define __constant__ +#define __shared__ +#define __constant__ #endif