From ef717b9a2ae60c1b8d9d81f098e6f04d21db0798 Mon Sep 17 00:00:00 2001 From: Laurent Morichetti Date: Mon, 23 May 2022 20:38:54 -0700 Subject: [PATCH] Add ROCtracer version information Change-Id: I10b268790d2dc4f3a3ad8624b2f553da6f3ccc8e [ROCm/roctracer commit: 1c450082af837b3c267eac0d4a32898063dd83ad] --- projects/roctracer/inc/ext/prof_protocol.h | 75 +++---- projects/roctracer/inc/roctracer.h | 207 ++++++++++-------- projects/roctracer/inc/roctracer_ext.h | 17 +- .../roctracer/src/roctracer/roctracer.cpp | 119 +++++----- .../roctracer/src/tracer_tool/tracer_tool.cpp | 73 +++--- .../test/app/MatrixTranspose_test.cpp | 63 +++--- 6 files changed, 297 insertions(+), 257 deletions(-) diff --git a/projects/roctracer/inc/ext/prof_protocol.h b/projects/roctracer/inc/ext/prof_protocol.h index f75e7c0dc2..d9aff1fb9d 100644 --- a/projects/roctracer/inc/ext/prof_protocol.h +++ b/projects/roctracer/inc/ext/prof_protocol.h @@ -18,80 +18,81 @@ OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE SOFTWARE. */ -#ifndef INC_EXT_PROF_PROTOCOL_H_ -#define INC_EXT_PROF_PROTOCOL_H_ +#ifndef EXT_PROF_PROTOCOL_H_ +#define EXT_PROF_PROTOCOL_H_ #include -// Traced API domains +/* Traced API domains */ typedef enum { - ACTIVITY_DOMAIN_HSA_API = 0, // HSA API domain - ACTIVITY_DOMAIN_HSA_OPS = 1, // HSA async activity domain - ACTIVITY_DOMAIN_HIP_OPS = 2, // HIP async activity domain - ACTIVITY_DOMAIN_HCC_OPS = ACTIVITY_DOMAIN_HIP_OPS, // HCC async activity domain - ACTIVITY_DOMAIN_HIP_VDI = ACTIVITY_DOMAIN_HIP_OPS, // HIP VDI async activity domain - ACTIVITY_DOMAIN_HIP_API = 3, // HIP API domain - ACTIVITY_DOMAIN_KFD_API = 4, // KFD API domain - ACTIVITY_DOMAIN_EXT_API = 5, // External ID domain - ACTIVITY_DOMAIN_ROCTX = 6, // ROCTX domain - ACTIVITY_DOMAIN_HSA_EVT = 7, // HSA events + ACTIVITY_DOMAIN_HSA_API = 0, /* HSA API domain */ + ACTIVITY_DOMAIN_HSA_OPS = 1, /* HSA async activity domain */ + ACTIVITY_DOMAIN_HIP_OPS = 2, /* HIP async activity domain */ + ACTIVITY_DOMAIN_HCC_OPS = ACTIVITY_DOMAIN_HIP_OPS, /* HCC async activity domain */ + ACTIVITY_DOMAIN_HIP_VDI = ACTIVITY_DOMAIN_HIP_OPS, /* HIP VDI async activity domain */ + ACTIVITY_DOMAIN_HIP_API = 3, /* HIP API domain */ + ACTIVITY_DOMAIN_KFD_API = 4, /* KFD API domain */ + ACTIVITY_DOMAIN_EXT_API = 5, /* External ID domain */ + ACTIVITY_DOMAIN_ROCTX = 6, /* ROCTX domain */ + ACTIVITY_DOMAIN_HSA_EVT = 7, /* HSA events */ ACTIVITY_DOMAIN_NUMBER } activity_domain_t; -// Extension API opcodes +/* Extension API opcodes */ typedef enum { ACTIVITY_EXT_OP_MARK = 0, ACTIVITY_EXT_OP_EXTERN_ID = 1 } activity_ext_op_t; -// API calback type +/* API callback type */ typedef void (*activity_rtapi_callback_t)(uint32_t domain, uint32_t cid, const void* data, void* arg); typedef uint32_t activity_kind_t; typedef uint32_t activity_op_t; -// API callback phase +/* API callback phase */ typedef enum { ACTIVITY_API_PHASE_ENTER = 0, ACTIVITY_API_PHASE_EXIT = 1 } activity_api_phase_t; -// Trace record types -// Correlation id +/* Trace record types */ + +/* Correlation id */ typedef uint64_t activity_correlation_id_t; -// Activity record type +/* Activity record type */ typedef struct activity_record_s { - uint32_t domain; // activity domain id - activity_kind_t kind; // activity kind - activity_op_t op; // activity op + uint32_t domain; /* activity domain id */ + activity_kind_t kind; /* activity kind */ + activity_op_t op; /* activity op */ union { struct { - activity_correlation_id_t correlation_id; // activity ID - uint64_t begin_ns; // host begin timestamp - uint64_t end_ns; // host end timestamp + activity_correlation_id_t correlation_id; /* activity ID */ + uint64_t begin_ns; /* host begin timestamp */ + uint64_t end_ns; /* host end timestamp */ }; struct { - uint32_t se; // sampled SE - uint64_t cycle; // sample cycle - uint64_t pc; // sample PC + uint32_t se; /* sampled SE */ + uint64_t cycle; /* sample cycle */ + uint64_t pc; /* sample PC */ } pc_sample; }; union { struct { - int device_id; // device id - uint64_t queue_id; // queue id + int device_id; /* device id */ + uint64_t queue_id; /* queue id */ }; struct { - uint32_t process_id; // device id - uint32_t thread_id; // thread id + uint32_t process_id; /* device id */ + uint32_t thread_id; /* thread id */ }; struct { - activity_correlation_id_t external_id; // external correlatino id + activity_correlation_id_t external_id; /* external correlation id */ }; }; - size_t bytes; // data size bytes + size_t bytes; /* data size bytes */ } activity_record_t; -// Activity sync calback type +/* Activity sync callback type */ typedef void* (*activity_sync_callback_t)(uint32_t cid, activity_record_t* record, const void* data, void* arg); -// Activity async calback type +/* Activity async callback type */ typedef void (*activity_id_callback_t)(activity_correlation_id_t id); typedef void (*activity_async_callback_t)(uint32_t op, void* record, void* arg); -#endif // INC_EXT_PROF_PROTOCOL_H_ +#endif /* EXT_PROF_PROTOCOL_H_ */ diff --git a/projects/roctracer/inc/roctracer.h b/projects/roctracer/inc/roctracer.h index 830572d2b4..8287d77656 100644 --- a/projects/roctracer/inc/roctracer.h +++ b/projects/roctracer/inc/roctracer.h @@ -29,27 +29,79 @@ * * The API provides functionality for registering the runtimes API callbacks and * asynchronous activity records pool support. -*/ + */ /** * \file * ROC tracer API interface. */ -#ifndef INC_ROCTRACER_H_ -#define INC_ROCTRACER_H_ +#ifndef ROCTRACER_H_ +#define ROCTRACER_H_ + +/* Placeholder for calling convention and import/export macros */ +#if !defined(ROCTRACER_CALL) +#define ROCTRACER_CALL +#endif /* !defined (ROCTRACER_CALL) */ + +#if !defined(ROCTRACER_EXPORT_DECORATOR) +#if defined(__GNUC__) +#define ROCTRACER_EXPORT_DECORATOR __attribute__((visibility("default"))) +#elif defined(_MSC_VER) +#define ROCTRACER_EXPORT_DECORATOR __declspec(dllexport) +#endif /* defined (_MSC_VER) */ +#endif /* !defined (ROCTRACER_EXPORT_DECORATOR) */ + +#if !defined(ROCTRACER_IMPORT_DECORATOR) +#if defined(__GNUC__) +#define ROCTRACER_IMPORT_DECORATOR +#elif defined(_MSC_VER) +#define ROCTRACER_IMPORT_DECORATOR __declspec(dllimport) +#endif /* defined (_MSC_VER) */ +#endif /* !defined (ROCTRACER_IMPORT_DECORATOR) */ + +#define ROCTRACER_EXPORT ROCTRACER_EXPORT_DECORATOR ROCTRACER_CALL +#define ROCTRACER_IMPORT ROCTRACER_IMPORT_DECORATOR ROCTRACER_CALL + +#if !defined(ROCTRACER) +#if defined(ROCTRACER_EXPORTS) +#define ROCTRACER_API ROCTRACER_EXPORT +#else /* !defined (ROCTRACER_EXPORTS) */ +#define ROCTRACER_API ROCTRACER_IMPORT +#endif /* !defined (ROCTRACER_EXPORTS) */ +#endif /* !defined (ROCTRACER) */ #include #include -#ifndef __cplusplus -#include -#endif #include #ifdef __cplusplus extern "C" { -#endif // __cplusplus +#endif /* __cplusplus */ + +/** \defgroup symbol_versions_group Symbol Versions + * + * The names used for the shared library versioned symbols. + * + * Every function is annotated with one of the version macros defined in this + * section. Each macro specifies a corresponding symbol version string. After + * dynamically loading the shared library with \p dlopen, the address of each + * function can be obtained using \p dlvsym with the name of the function and + * its corresponding symbol version string. An error will be reported by \p + * dlvsym if the installed library does not support the version for the + * function specified in this version of the interface. + * + * @{ + */ + +/** + * The function was introduced in version 4.1 of the interface and has the + * symbol version string of ``"ROCTRACER_4.1"``. + */ +#define ROCTRACER_VERSION_4_1 + +/** @} */ /** \defgroup versioning_group Versioning * @@ -74,7 +126,7 @@ extern "C" { * The minor version of the interface as a macro so it can be used by the * preprocessor. */ -#define ROCTRACER_VERSION_MINOR 0 +#define ROCTRACER_VERSION_MINOR 1 /** * Query the major version of the installed library. @@ -83,7 +135,7 @@ extern "C" { * check if it is compatible with this interface version. This function can be * used even when the library is not initialized. */ -uint32_t roctracer_version_major(); +ROCTRACER_API uint32_t roctracer_version_major() ROCTRACER_VERSION_4_1; /** * Query the minor version of the installed library. @@ -92,7 +144,7 @@ uint32_t roctracer_version_major(); * check if it is compatible with this interface version. This function can be * used even when the library is not initialized. */ -uint32_t roctracer_version_minor(); +ROCTRACER_API uint32_t roctracer_version_minor() ROCTRACER_VERSION_4_1; /** @} */ @@ -188,7 +240,7 @@ typedef enum { * \return Return the error string. The caller owns the returned string and * should use \p free() to deallocate it. */ -const char* roctracer_error_string(); +ROCTRACER_API const char* roctracer_error_string() ROCTRACER_VERSION_4_1; /** @} */ @@ -218,9 +270,8 @@ typedef activity_domain_t roctracer_domain_t; * the domain or operation are invalid. The string is owned by the ROC Tracer * library. */ -const char* roctracer_op_string(uint32_t domain, - uint32_t op, - uint32_t kind); +ROCTRACER_API const char* roctracer_op_string(uint32_t domain, uint32_t op, + uint32_t kind) ROCTRACER_VERSION_4_1; /** * Query the operation code given a domain and the name of an operation. @@ -242,10 +293,8 @@ const char* roctracer_op_string(uint32_t domain, * @retval ::ROCTRACER_STATUS_ERROR_INVALID_DOMAIN_ID The domain is invalid or * not supported. */ -roctracer_status_t roctracer_op_code(uint32_t domain, - const char* str, - uint32_t* op, - uint32_t* kind); +ROCTRACER_API roctracer_status_t roctracer_op_code(uint32_t domain, const char* str, uint32_t* op, + uint32_t* kind) ROCTRACER_VERSION_4_1; /** * Set the properties of a domain. @@ -259,8 +308,8 @@ roctracer_status_t roctracer_op_code(uint32_t domain, * @retval ::ROCTRACER_STATUS_SUCCESS The function has been executed * successfully. */ -roctracer_status_t roctracer_set_properties(roctracer_domain_t domain, - void* properties); +ROCTRACER_API roctracer_status_t roctracer_set_properties(roctracer_domain_t domain, + void* properties) ROCTRACER_VERSION_4_1; /** @} */ @@ -301,11 +350,9 @@ typedef activity_rtapi_callback_t roctracer_rtapi_callback_t; * @retval ::ROCTRACER_STATUS_ERROR_INVALID_ARGUMENT \p op is invalid for \p * domain. */ -roctracer_status_t roctracer_enable_op_callback( - activity_domain_t domain, - uint32_t op, - activity_rtapi_callback_t callback, - void* arg); +ROCTRACER_API roctracer_status_t roctracer_enable_op_callback(activity_domain_t domain, uint32_t op, + activity_rtapi_callback_t callback, + void* arg) ROCTRACER_VERSION_4_1; /** * Enable runtime API callback for all operations of a domain. @@ -322,10 +369,8 @@ roctracer_status_t roctracer_enable_op_callback( * * @retval ::ROCTRACER_STATUS_ERROR_INVALID_DOMAIN_ID \p domain is invalid. */ -roctracer_status_t roctracer_enable_domain_callback( - activity_domain_t domain, - activity_rtapi_callback_t callback, - void* arg); +ROCTRACER_API roctracer_status_t roctracer_enable_domain_callback( + activity_domain_t domain, activity_rtapi_callback_t callback, void* arg) ROCTRACER_VERSION_4_1; /** * Enable runtime API callback for all operations of all domains. @@ -338,9 +383,8 @@ roctracer_status_t roctracer_enable_domain_callback( * @retval ::ROCTRACER_STATUS_SUCCESS The function has been executed * successfully. */ -roctracer_status_t roctracer_enable_callback( - activity_rtapi_callback_t callback, - void* arg); +ROCTRACER_API roctracer_status_t roctracer_enable_callback(activity_rtapi_callback_t callback, + void* arg) ROCTRACER_VERSION_4_1; /** * Disable runtime API callback for a specific operation of a domain. @@ -357,8 +401,8 @@ roctracer_status_t roctracer_enable_callback( * @retval ::ROCTRACER_STATUS_ERROR_INVALID_ARGUMENT \p op is invalid for \p * domain. */ -roctracer_status_t roctracer_disable_op_callback(activity_domain_t domain, - uint32_t op); +ROCTRACER_API roctracer_status_t roctracer_disable_op_callback(activity_domain_t domain, + uint32_t op) ROCTRACER_VERSION_4_1; /** * Disable runtime API callback for all operations of a domain. @@ -370,7 +414,8 @@ roctracer_status_t roctracer_disable_op_callback(activity_domain_t domain, * * @retval ::ROCTRACER_STATUS_ERROR_INVALID_DOMAIN_ID \p domain is invalid. */ -roctracer_status_t roctracer_disable_domain_callback(activity_domain_t domain); +ROCTRACER_API roctracer_status_t roctracer_disable_domain_callback(activity_domain_t domain) + ROCTRACER_VERSION_4_1; /** * Disable runtime API callback for all operations of all domains. @@ -378,7 +423,7 @@ roctracer_status_t roctracer_disable_domain_callback(activity_domain_t domain); * @retval ::ROCTRACER_STATUS_SUCCESS The function has been executed * successfully. */ -roctracer_status_t roctracer_disable_callback(); +ROCTRACER_API roctracer_status_t roctracer_disable_callback() ROCTRACER_VERSION_4_1; /** @} */ @@ -413,20 +458,19 @@ typedef activity_record_t roctracer_record_t; * @retval ::ROCTRACER_STATUS_SUCCESS The function has been executed * successfully. */ -roctracer_status_t roctracer_next_record( - const activity_record_t* record, - const activity_record_t** next); +ROCTRACER_API roctracer_status_t roctracer_next_record( + const activity_record_t* record, const activity_record_t** next) ROCTRACER_VERSION_4_1; /** * Memory pool allocator callback. * * If \p *ptr is NULL, then allocate memory of \p size bytes and save address * in \p *ptr. - * + * * If \p *ptr is non-NULL and size is non-0, then reallocate the memory at \p * *ptr with size \p size and save the address in \p *ptr. The memory will have * been allocated by the same callback. - * + * * If \p *ptr is non-NULL and size is 0, then deallocate the memory at \p *ptr. * The memory will have been allocated by the same callback. * @@ -436,9 +480,7 @@ roctracer_status_t roctracer_next_record( * \p arg Argument provided in the ::roctracer_properties_t passed to the * ::roctracer_open_pool function. */ -typedef void (*roctracer_allocator_t)(char** ptr, - size_t size, - void* arg); +typedef void (*roctracer_allocator_t)(char** ptr, size_t size, void* arg); /** * Memory pool buffer callback. @@ -452,10 +494,7 @@ typedef void (*roctracer_allocator_t)(char** ptr, * * \p arg the argument specified when the callback was defined. */ -typedef void (*roctracer_buffer_callback_t)( - const char* begin, // [in] available buffered trace records - const char* end, // [in] end of buffered trace records - void* arg); // [in/out] callback arg +typedef void (*roctracer_buffer_callback_t)(const char* begin, const char* end, void* arg); /** * Memory pool properties. @@ -463,7 +502,6 @@ typedef void (*roctracer_buffer_callback_t)( * Defines the propertis when a tracer memory pool is created. */ typedef struct { - /** * ROC Tracer mode. */ @@ -523,9 +561,8 @@ typedef void roctracer_pool_t; * @retval ROCTRACER_STATUS_ERROR_MEMORY_ALLOCATION Unable to allocate memory * for the \p pool. Unable to create the pool. */ -roctracer_status_t roctracer_open_pool_expl( - const roctracer_properties_t* properties, - roctracer_pool_t** pool); +ROCTRACER_API roctracer_status_t roctracer_open_pool_expl( + const roctracer_properties_t* properties, roctracer_pool_t** pool) ROCTRACER_VERSION_4_1; /** * Create tracer memory pool. @@ -544,7 +581,8 @@ roctracer_status_t roctracer_open_pool_expl( * @retval ROCTRACER_STATUS_ERROR_MEMORY_ALLOCATION Unable to allocate memory * for the \p pool. Unable to create the pool. */ -roctracer_status_t roctracer_open_pool(const roctracer_properties_t* properties); +ROCTRACER_API roctracer_status_t roctracer_open_pool(const roctracer_properties_t* properties) + ROCTRACER_VERSION_4_1; /** * Close tracer memory pool. @@ -559,8 +597,8 @@ roctracer_status_t roctracer_open_pool(const roctracer_properties_t* properties) * @retval ::ROCTRACER_STATUS_SUCCESS The function has been executed * successfully or pool was NULL and there is no default pool. */ -roctracer_status_t roctracer_close_pool_expl( - roctracer_pool_t* pool); +ROCTRACER_API roctracer_status_t roctracer_close_pool_expl(roctracer_pool_t* pool) + ROCTRACER_VERSION_4_1; /** * Close default tracer memory pool, if defined, and set to undefined. @@ -572,7 +610,7 @@ roctracer_status_t roctracer_close_pool_expl( * @retval ::ROCTRACER_STATUS_SUCCESS The function has been executed * successfully or there is no default pool. */ -roctracer_status_t roctracer_close_pool(); +ROCTRACER_API roctracer_status_t roctracer_close_pool() ROCTRACER_VERSION_4_1; /** * Query and set the default memory pool. @@ -583,17 +621,15 @@ roctracer_status_t roctracer_close_pool(); * @return Return the current default memory pool before any change, or NULL if * none is defined. */ -// Return current default pool -// Set new default pool if the argument is not NULL -roctracer_pool_t* roctracer_default_pool_expl( - roctracer_pool_t* pool); +ROCTRACER_API roctracer_pool_t* roctracer_default_pool_expl(roctracer_pool_t* pool) + ROCTRACER_VERSION_4_1; /** * Query the current default memory pool. * * @return Return the current default memory pool, or NULL is none is defined. */ -roctracer_pool_t* roctracer_default_pool(); +ROCTRACER_API roctracer_pool_t* roctracer_default_pool() ROCTRACER_VERSION_4_1; /** * Enable activity record logging for a specified operation of a domain @@ -611,10 +647,8 @@ roctracer_pool_t* roctracer_default_pool(); * * @retval ROCTRACER_STATUS_ERROR \p pool is NULL and no default pool is defined. */ -roctracer_status_t roctracer_enable_op_activity_expl( - activity_domain_t domain, - uint32_t op, - roctracer_pool_t* pool); +ROCTRACER_API roctracer_status_t roctracer_enable_op_activity_expl( + activity_domain_t domain, uint32_t op, roctracer_pool_t* pool) ROCTRACER_VERSION_4_1; /** * Enable activity record logging for a specified operation of a domain using @@ -629,9 +663,8 @@ roctracer_status_t roctracer_enable_op_activity_expl( * * @retval ROCTRACER_STATUS_ERROR No default pool is defined. */ -roctracer_status_t roctracer_enable_op_activity( - activity_domain_t domain, - uint32_t op); +ROCTRACER_API roctracer_status_t roctracer_enable_op_activity(activity_domain_t domain, + uint32_t op) ROCTRACER_VERSION_4_1; /** * Enable activity record logging for all operations of a domain providing a @@ -647,9 +680,8 @@ roctracer_status_t roctracer_enable_op_activity( * * @retval ROCTRACER_STATUS_ERROR \p pool is NULL and no default pool is defined. */ -roctracer_status_t roctracer_enable_domain_activity_expl( - activity_domain_t domain, - roctracer_pool_t* pool); +ROCTRACER_API roctracer_status_t roctracer_enable_domain_activity_expl( + activity_domain_t domain, roctracer_pool_t* pool) ROCTRACER_VERSION_4_1; /** * Enable activity record logging for all operations of a domain using the @@ -662,7 +694,8 @@ roctracer_status_t roctracer_enable_domain_activity_expl( * * @retval ROCTRACER_STATUS_ERROR No default pool is defined. */ -roctracer_status_t roctracer_enable_domain_activity(activity_domain_t domain); +ROCTRACER_API roctracer_status_t roctracer_enable_domain_activity(activity_domain_t domain) + ROCTRACER_VERSION_4_1; /** * Enable activity record logging for all operations of all domains providing a @@ -676,8 +709,8 @@ roctracer_status_t roctracer_enable_domain_activity(activity_domain_t domain); * * @retval ROCTRACER_STATUS_ERROR \p pool is NULL and no default pool is defined. */ -roctracer_status_t roctracer_enable_activity_expl( - roctracer_pool_t* pool); +ROCTRACER_API roctracer_status_t roctracer_enable_activity_expl(roctracer_pool_t* pool) + ROCTRACER_VERSION_4_1; /** * Enable activity record logging for all operations of all domains using the @@ -688,7 +721,7 @@ roctracer_status_t roctracer_enable_activity_expl( * * @retval ROCTRACER_STATUS_ERROR No default pool is defined. */ -roctracer_status_t roctracer_enable_activity(); +ROCTRACER_API roctracer_status_t roctracer_enable_activity() ROCTRACER_VERSION_4_1; /** * Disable activity record logging for a specified operation of a domain. @@ -700,9 +733,8 @@ roctracer_status_t roctracer_enable_activity(); * @retval ::ROCTRACER_STATUS_SUCCESS The function has been executed * successfully. */ -// -roctracer_status_t roctracer_disable_op_activity(activity_domain_t domain, - uint32_t op); +ROCTRACER_API roctracer_status_t roctracer_disable_op_activity(activity_domain_t domain, + uint32_t op) ROCTRACER_VERSION_4_1; /** * Disable activity record logging for all operations of a domain. @@ -712,7 +744,8 @@ roctracer_status_t roctracer_disable_op_activity(activity_domain_t domain, * @retval ::ROCTRACER_STATUS_SUCCESS The function has been executed * successfully. */ -roctracer_status_t roctracer_disable_domain_activity(activity_domain_t domain); +ROCTRACER_API roctracer_status_t roctracer_disable_domain_activity(activity_domain_t domain) + ROCTRACER_VERSION_4_1; /** * Disable activity record logging for all operations of all domains. @@ -722,7 +755,7 @@ roctracer_status_t roctracer_disable_domain_activity(activity_domain_t domain); * @retval ::ROCTRACER_STATUS_SUCCESS The function has been executed * successfully. */ -roctracer_status_t roctracer_disable_activity(); +ROCTRACER_API roctracer_status_t roctracer_disable_activity() ROCTRACER_VERSION_4_1; /** * Flush available activity records for a memory pool. @@ -737,8 +770,8 @@ roctracer_status_t roctracer_disable_activity(); * @retval ::ROCTRACER_STATUS_SUCCESS The function has been executed * successfully. */ -roctracer_status_t roctracer_flush_activity_expl( - roctracer_pool_t* pool); +ROCTRACER_API roctracer_status_t roctracer_flush_activity_expl(roctracer_pool_t* pool) + ROCTRACER_VERSION_4_1; /** * Flush available activity records for the default memory pool. @@ -750,7 +783,7 @@ roctracer_status_t roctracer_flush_activity_expl( * @retval ::ROCTRACER_STATUS_SUCCESS The function has been executed * successfully. */ -roctracer_status_t roctracer_flush_activity(); +ROCTRACER_API roctracer_status_t roctracer_flush_activity() ROCTRACER_VERSION_4_1; /** @} */ @@ -769,12 +802,12 @@ roctracer_status_t roctracer_flush_activity(); * @retval ::ROCTRACER_STATUS_SUCCESS The function has been executed * successfully. */ -roctracer_status_t roctracer_get_timestamp(uint64_t* timestamp); +ROCTRACER_API roctracer_status_t roctracer_get_timestamp(uint64_t* timestamp) ROCTRACER_VERSION_4_1; /** @} */ #ifdef __cplusplus -} // extern "C" block -#endif // __cplusplus +} /* extern "C" block */ +#endif /* __cplusplus */ -#endif // INC_ROCTRACER_H_ +#endif /* ROCTRACER_H_ */ diff --git a/projects/roctracer/inc/roctracer_ext.h b/projects/roctracer/inc/roctracer_ext.h index ccd56617a3..30c30136ea 100644 --- a/projects/roctracer/inc/roctracer_ext.h +++ b/projects/roctracer/inc/roctracer_ext.h @@ -27,8 +27,8 @@ // //////////////////////////////////////////////////////////////////////////////// -#ifndef INC_ROCTRACER_EXT_H_ -#define INC_ROCTRACER_EXT_H_ +#ifndef ROCTRACER_EXT_H_ +#define ROCTRACER_EXT_H_ #include @@ -47,26 +47,27 @@ extern "C" { // Application annotation API // Tracing start API -void roctracer_start(); +void ROCTRACER_API roctracer_start() ROCTRACER_VERSION_4_1; // Tracing stop API -void roctracer_stop(); +void ROCTRACER_API roctracer_stop() ROCTRACER_VERSION_4_1; //////////////////////////////////////////////////////////////////////////////// // External correlation id API // Notifies that the calling thread is entering an external API region. // Push an external correlation id for the calling thread. -roctracer_status_t roctracer_activity_push_external_correlation_id(activity_correlation_id_t id); +roctracer_status_t ROCTRACER_API +roctracer_activity_push_external_correlation_id(activity_correlation_id_t id) ROCTRACER_VERSION_4_1; // Notifies that the calling thread is leaving an external API region. // Pop an external correlation id for the calling thread. // 'lastId' returns the last external correlation if not NULL -roctracer_status_t roctracer_activity_pop_external_correlation_id( - activity_correlation_id_t* last_id); +roctracer_status_t ROCTRACER_API roctracer_activity_pop_external_correlation_id( + activity_correlation_id_t* last_id) ROCTRACER_VERSION_4_1; #ifdef __cplusplus } // extern "C" block #endif // __cplusplus -#endif // INC_ROCTRACER_EXT_H_ +#endif // ROCTRACER_EXT_H_ diff --git a/projects/roctracer/src/roctracer/roctracer.cpp b/projects/roctracer/src/roctracer/roctracer.cpp index 655bc48b48..5e132382fd 100644 --- a/projects/roctracer/src/roctracer/roctracer.cpp +++ b/projects/roctracer/src/roctracer/roctracer.cpp @@ -43,10 +43,6 @@ #include "exception.h" #include "util/logger.h" -#define PUBLIC_API __attribute__((visibility("default"))) -#define CONSTRUCTOR_API __attribute__((constructor)) -#define DESTRUCTOR_API __attribute__((destructor)) - #define CHECK_HSA_STATUS(msg, status) \ do { \ if ((status) != HSA_STATUS_SUCCESS) { \ @@ -520,17 +516,17 @@ LOADER_INSTANTIATE(); // // Returns library version -PUBLIC_API uint32_t roctracer_version_major() { return ROCTRACER_VERSION_MAJOR; } -PUBLIC_API uint32_t roctracer_version_minor() { return ROCTRACER_VERSION_MINOR; } +ROCTRACER_API uint32_t roctracer_version_major() { return ROCTRACER_VERSION_MAJOR; } +ROCTRACER_API uint32_t roctracer_version_minor() { return ROCTRACER_VERSION_MINOR; } // Returns the last error -PUBLIC_API const char* roctracer_error_string() { +ROCTRACER_API const char* roctracer_error_string() { return strdup(util::Logger::LastMessage().c_str()); } // Return Op string by given domain and activity/API codes // nullptr returned on the error and the library errno is set -PUBLIC_API const char* roctracer_op_string(uint32_t domain, uint32_t op, uint32_t kind) { +ROCTRACER_API const char* roctracer_op_string(uint32_t domain, uint32_t op, uint32_t kind) { API_METHOD_PREFIX switch (domain) { case ACTIVITY_DOMAIN_HSA_API: @@ -552,8 +548,8 @@ PUBLIC_API const char* roctracer_op_string(uint32_t domain, uint32_t op, uint32_ } // Return Op code and kind by given string -PUBLIC_API roctracer_status_t roctracer_op_code(uint32_t domain, const char* str, uint32_t* op, - uint32_t* kind) { +ROCTRACER_API roctracer_status_t roctracer_op_code(uint32_t domain, const char* str, uint32_t* op, + uint32_t* kind) { API_METHOD_PREFIX switch (domain) { case ACTIVITY_DOMAIN_HSA_API: { @@ -687,17 +683,17 @@ static void roctracer_enable_callback_impl(roctracer_domain_t domain, uint32_t o roctracer_enable_callback_fun(domain, op, callback, user_data); } -PUBLIC_API roctracer_status_t roctracer_enable_op_callback(roctracer_domain_t domain, uint32_t op, - roctracer_rtapi_callback_t callback, - void* user_data) { +ROCTRACER_API roctracer_status_t roctracer_enable_op_callback(roctracer_domain_t domain, + uint32_t op, + roctracer_rtapi_callback_t callback, + void* user_data) { API_METHOD_PREFIX roctracer_enable_callback_impl(domain, op, callback, user_data); API_METHOD_SUFFIX } -PUBLIC_API roctracer_status_t roctracer_enable_domain_callback(roctracer_domain_t domain, - roctracer_rtapi_callback_t callback, - void* user_data) { +ROCTRACER_API roctracer_status_t roctracer_enable_domain_callback( + roctracer_domain_t domain, roctracer_rtapi_callback_t callback, void* user_data) { API_METHOD_PREFIX const uint32_t op_end = get_op_end(domain); for (uint32_t op = get_op_begin(domain); op < op_end; ++op) @@ -705,8 +701,8 @@ PUBLIC_API roctracer_status_t roctracer_enable_domain_callback(roctracer_domain_ API_METHOD_SUFFIX } -PUBLIC_API roctracer_status_t roctracer_enable_callback(roctracer_rtapi_callback_t callback, - void* user_data) { +ROCTRACER_API roctracer_status_t roctracer_enable_callback(roctracer_rtapi_callback_t callback, + void* user_data) { API_METHOD_PREFIX for (uint32_t domain = 0; domain < ACTIVITY_DOMAIN_NUMBER; ++domain) { const uint32_t op_end = get_op_end(domain); @@ -771,14 +767,14 @@ static void roctracer_disable_callback_impl(roctracer_domain_t domain, uint32_t roctracer_disable_callback_fun(domain, op); } -PUBLIC_API roctracer_status_t roctracer_disable_op_callback(roctracer_domain_t domain, - uint32_t op) { +ROCTRACER_API roctracer_status_t roctracer_disable_op_callback(roctracer_domain_t domain, + uint32_t op) { API_METHOD_PREFIX roctracer_disable_callback_impl(domain, op); API_METHOD_SUFFIX } -PUBLIC_API roctracer_status_t roctracer_disable_domain_callback(roctracer_domain_t domain) { +ROCTRACER_API roctracer_status_t roctracer_disable_domain_callback(roctracer_domain_t domain) { API_METHOD_PREFIX const uint32_t op_end = get_op_end(domain); for (uint32_t op = get_op_begin(domain); op < op_end; ++op) @@ -786,7 +782,7 @@ PUBLIC_API roctracer_status_t roctracer_disable_domain_callback(roctracer_domain API_METHOD_SUFFIX } -PUBLIC_API roctracer_status_t roctracer_disable_callback() { +ROCTRACER_API roctracer_status_t roctracer_disable_callback() { API_METHOD_PREFIX for (uint32_t domain = 0; domain < ACTIVITY_DOMAIN_NUMBER; ++domain) { const uint32_t op_end = get_op_end(domain); @@ -797,14 +793,14 @@ PUBLIC_API roctracer_status_t roctracer_disable_callback() { } // Return default pool and set new one if parameter pool is not NULL. -PUBLIC_API roctracer_pool_t* roctracer_default_pool_expl(roctracer_pool_t* pool) { +ROCTRACER_API roctracer_pool_t* roctracer_default_pool_expl(roctracer_pool_t* pool) { std::lock_guard lock(memory_pool_mutex); roctracer_pool_t* p = reinterpret_cast(default_memory_pool); if (pool != nullptr) default_memory_pool = reinterpret_cast(pool); return p; } -PUBLIC_API roctracer_pool_t* roctracer_default_pool() { +ROCTRACER_API roctracer_pool_t* roctracer_default_pool() { std::lock_guard lock(memory_pool_mutex); return reinterpret_cast(default_memory_pool); } @@ -824,21 +820,21 @@ static void roctracer_open_pool_impl(const roctracer_properties_t* properties, default_memory_pool = p; } -PUBLIC_API roctracer_status_t roctracer_open_pool_expl(const roctracer_properties_t* properties, - roctracer_pool_t** pool) { +ROCTRACER_API roctracer_status_t roctracer_open_pool_expl(const roctracer_properties_t* properties, + roctracer_pool_t** pool) { API_METHOD_PREFIX roctracer_open_pool_impl(properties, pool); API_METHOD_SUFFIX } -PUBLIC_API roctracer_status_t roctracer_open_pool(const roctracer_properties_t* properties) { +ROCTRACER_API roctracer_status_t roctracer_open_pool(const roctracer_properties_t* properties) { API_METHOD_PREFIX roctracer_open_pool_impl(properties, nullptr); API_METHOD_SUFFIX } -PUBLIC_API roctracer_status_t roctracer_next_record(const activity_record_t* record, - const activity_record_t** next) { +ROCTRACER_API roctracer_status_t roctracer_next_record(const activity_record_t* record, + const activity_record_t** next) { API_METHOD_PREFIX *next = record + 1; API_METHOD_SUFFIX @@ -912,15 +908,16 @@ static void roctracer_enable_activity_impl(roctracer_domain_t domain, uint32_t o roctracer_enable_activity_fun(domain, op, pool); } -PUBLIC_API roctracer_status_t roctracer_enable_op_activity_expl(roctracer_domain_t domain, - uint32_t op, - roctracer_pool_t* pool) { +ROCTRACER_API roctracer_status_t roctracer_enable_op_activity_expl(roctracer_domain_t domain, + uint32_t op, + roctracer_pool_t* pool) { API_METHOD_PREFIX roctracer_enable_activity_impl(domain, op, pool); API_METHOD_SUFFIX } -PUBLIC_API roctracer_status_t roctracer_enable_op_activity(activity_domain_t domain, uint32_t op) { +ROCTRACER_API roctracer_status_t roctracer_enable_op_activity(activity_domain_t domain, + uint32_t op) { API_METHOD_PREFIX roctracer_enable_activity_impl(domain, op, nullptr); API_METHOD_SUFFIX @@ -933,14 +930,14 @@ static void roctracer_enable_domain_activity_impl(roctracer_domain_t domain, roctracer_enable_activity_impl(domain, op, pool); } -PUBLIC_API roctracer_status_t roctracer_enable_domain_activity_expl(roctracer_domain_t domain, - roctracer_pool_t* pool) { +ROCTRACER_API roctracer_status_t roctracer_enable_domain_activity_expl(roctracer_domain_t domain, + roctracer_pool_t* pool) { API_METHOD_PREFIX roctracer_enable_domain_activity_impl(domain, pool); API_METHOD_SUFFIX } -PUBLIC_API roctracer_status_t roctracer_enable_domain_activity(activity_domain_t domain) { +ROCTRACER_API roctracer_status_t roctracer_enable_domain_activity(activity_domain_t domain) { API_METHOD_PREFIX roctracer_enable_domain_activity_impl(domain, nullptr); API_METHOD_SUFFIX @@ -954,13 +951,13 @@ static void roctracer_enable_activity_impl(roctracer_pool_t* pool) { } } -PUBLIC_API roctracer_status_t roctracer_enable_activity_expl(roctracer_pool_t* pool) { +ROCTRACER_API roctracer_status_t roctracer_enable_activity_expl(roctracer_pool_t* pool) { API_METHOD_PREFIX roctracer_enable_activity_impl(pool); API_METHOD_SUFFIX } -PUBLIC_API roctracer_status_t roctracer_enable_activity() { +ROCTRACER_API roctracer_status_t roctracer_enable_activity() { API_METHOD_PREFIX roctracer_enable_activity_impl(nullptr); API_METHOD_SUFFIX @@ -1019,14 +1016,14 @@ static void roctracer_disable_activity_impl(roctracer_domain_t domain, uint32_t roctracer_disable_activity_fun(domain, op); } -PUBLIC_API roctracer_status_t roctracer_disable_op_activity(roctracer_domain_t domain, - uint32_t op) { +ROCTRACER_API roctracer_status_t roctracer_disable_op_activity(roctracer_domain_t domain, + uint32_t op) { API_METHOD_PREFIX roctracer_disable_activity_impl(domain, op); API_METHOD_SUFFIX } -PUBLIC_API roctracer_status_t roctracer_disable_domain_activity(roctracer_domain_t domain) { +ROCTRACER_API roctracer_status_t roctracer_disable_domain_activity(roctracer_domain_t domain) { API_METHOD_PREFIX const uint32_t op_end = get_op_end(domain); for (uint32_t op = get_op_begin(domain); op < op_end; ++op) @@ -1034,7 +1031,7 @@ PUBLIC_API roctracer_status_t roctracer_disable_domain_activity(roctracer_domain API_METHOD_SUFFIX } -PUBLIC_API roctracer_status_t roctracer_disable_activity() { +ROCTRACER_API roctracer_status_t roctracer_disable_activity() { API_METHOD_PREFIX for (uint32_t domain = 0; domain < ACTIVITY_DOMAIN_NUMBER; ++domain) { const uint32_t op_end = get_op_end(domain); @@ -1064,13 +1061,13 @@ static void roctracer_close_pool_impl(roctracer_pool_t* pool) { delete (p); } -PUBLIC_API roctracer_status_t roctracer_close_pool_expl(roctracer_pool_t* pool) { +ROCTRACER_API roctracer_status_t roctracer_close_pool_expl(roctracer_pool_t* pool) { API_METHOD_PREFIX roctracer_close_pool_impl(pool); API_METHOD_SUFFIX } -PUBLIC_API roctracer_status_t roctracer_close_pool() { +ROCTRACER_API roctracer_status_t roctracer_close_pool() { API_METHOD_PREFIX roctracer_close_pool_impl(NULL); API_METHOD_SUFFIX @@ -1083,13 +1080,13 @@ static void roctracer_flush_activity_impl(roctracer_pool_t* pool) { if (default_memory_pool != nullptr) default_memory_pool->Flush(); } -PUBLIC_API roctracer_status_t roctracer_flush_activity_expl(roctracer_pool_t* pool) { +ROCTRACER_API roctracer_status_t roctracer_flush_activity_expl(roctracer_pool_t* pool) { API_METHOD_PREFIX roctracer_flush_activity_impl(pool); API_METHOD_SUFFIX } -PUBLIC_API roctracer_status_t roctracer_flush_activity() { +ROCTRACER_API roctracer_status_t roctracer_flush_activity() { API_METHOD_PREFIX roctracer_flush_activity_impl(nullptr); API_METHOD_SUFFIX @@ -1097,7 +1094,7 @@ PUBLIC_API roctracer_status_t roctracer_flush_activity() { // Notifies that the calling thread is entering an external API region. // Push an external correlation id for the calling thread. -PUBLIC_API roctracer_status_t +ROCTRACER_API roctracer_status_t roctracer_activity_push_external_correlation_id(activity_correlation_id_t id) { API_METHOD_PREFIX external_id_stack.push(id); @@ -1107,7 +1104,7 @@ roctracer_activity_push_external_correlation_id(activity_correlation_id_t id) { // Notifies that the calling thread is leaving an external API region. // Pop an external correlation id for the calling thread. // 'lastId' returns the last external correlation -PUBLIC_API roctracer_status_t +ROCTRACER_API roctracer_status_t roctracer_activity_pop_external_correlation_id(activity_correlation_id_t* last_id) { API_METHOD_PREFIX if (last_id != nullptr) *last_id = 0; @@ -1119,8 +1116,8 @@ roctracer_activity_pop_external_correlation_id(activity_correlation_id_t* last_i API_METHOD_SUFFIX } -// Mark API -extern "C" PUBLIC_API void roctracer_mark(const char* str) { +// Mark API (FIXME: why isn't it in the roctracer_ext.h header?) +extern "C" ROCTRACER_API void roctracer_mark(const char* str) { if (mark_api_callback_ptr) { mark_api_callback_ptr(ACTIVITY_DOMAIN_EXT_API, ACTIVITY_EXT_OP_MARK, str, nullptr); NextCorrelationId(); // account for user-defined markers when tracking @@ -1129,7 +1126,7 @@ extern "C" PUBLIC_API void roctracer_mark(const char* str) { } // Start API -PUBLIC_API void roctracer_start() { +ROCTRACER_API void roctracer_start() { if (set_stopped(0)) { if (ext_support::roctracer_start_cb) ext_support::roctracer_start_cb(); cb_journal.ForEach([](roctracer_domain_t domain, uint32_t op, const CallbackJournalData& data) { @@ -1145,7 +1142,7 @@ PUBLIC_API void roctracer_start() { } // Stop API -PUBLIC_API void roctracer_stop() { +ROCTRACER_API void roctracer_stop() { if (set_stopped(1)) { // Must disable the activity first as the spawner checks for the activity being NULL // to indicate that there is no callback. @@ -1161,15 +1158,15 @@ PUBLIC_API void roctracer_stop() { } } -PUBLIC_API roctracer_status_t roctracer_get_timestamp(uint64_t* timestamp) { +ROCTRACER_API roctracer_status_t roctracer_get_timestamp(uint64_t* timestamp) { API_METHOD_PREFIX *timestamp = util::timestamp_ns(); API_METHOD_SUFFIX } // Set properties -PUBLIC_API roctracer_status_t roctracer_set_properties(roctracer_domain_t domain, - void* properties) { +ROCTRACER_API roctracer_status_t roctracer_set_properties(roctracer_domain_t domain, + void* properties) { API_METHOD_PREFIX switch (domain) { case ACTIVITY_DOMAIN_HSA_OPS: { @@ -1219,24 +1216,24 @@ PUBLIC_API roctracer_status_t roctracer_set_properties(roctracer_domain_t domain API_METHOD_SUFFIX } -CONSTRUCTOR_API void constructor() { +__attribute__((constructor)) void constructor() { ONLOAD_TRACE_BEG(); util::Logger::Create(); ONLOAD_TRACE_END(); } -DESTRUCTOR_API void destructor() { +__attribute__((destructor)) void destructor() { ONLOAD_TRACE_BEG(); util::Logger::Destroy(); ONLOAD_TRACE_END(); } // HSA-runtime tool on-load method -extern "C" PUBLIC_API bool OnLoad(HsaApiTable* table, uint64_t runtime_version, - uint64_t failed_tool_count, - const char* const* failed_tool_names) { +extern "C" ROCTRACER_EXPORT bool OnLoad(HsaApiTable* table, uint64_t runtime_version, + uint64_t failed_tool_count, + const char* const* failed_tool_names) { hsa_support::SaveHsaApi(table); return true; } -extern "C" PUBLIC_API void OnUnload() {} \ No newline at end of file +extern "C" ROCTRACER_EXPORT void OnUnload() {} \ No newline at end of file diff --git a/projects/roctracer/src/tracer_tool/tracer_tool.cpp b/projects/roctracer/src/tracer_tool/tracer_tool.cpp index 78fefc5f73..b49dce2788 100644 --- a/projects/roctracer/src/tracer_tool/tracer_tool.cpp +++ b/projects/roctracer/src/tracer_tool/tracer_tool.cpp @@ -27,7 +27,6 @@ #include /* names denangle */ #include -#include #include #include #include @@ -46,23 +45,19 @@ #include "trace_buffer.h" #include "evt_stats.h" -#define PUBLIC_API __attribute__((visibility("default"))) #define CONSTRUCTOR_API __attribute__((constructor)) #define DESTRUCTOR_API __attribute__((destructor)) -// Linux sys call -#define PTHREAD_CALL(call) \ - do { \ - int err = call; \ - if (err != 0) { \ - errno = err; \ - perror(#call); \ - abort(); \ - } \ - } while (0) +#if !defined(ROCTRACER_TOOL_EXPORT) +#if defined(__GNUC__) +#define ROCTRACER_TOOL_EXPORT __attribute__((visibility("default"))) +#elif defined(_MSC_VER) +#define ROCTRACER_TOOL_EXPORT __declspec(dllexport) +#endif /* defined (_MSC_VER) */ +#endif /* !defined (ROCTRACER_TOOL_EXPORT) */ // Macro to check ROC-tracer calls status -#define ROCTRACER_CALL(call) \ +#define CHECK_ROCTRACER(call) \ do { \ int err = call; \ if (err != 0) { \ @@ -107,7 +102,7 @@ namespace util { inline timestamp_t timestamp_ns() { timestamp_t timestamp; - ROCTRACER_CALL(roctracer_get_timestamp(×tamp)); + CHECK_ROCTRACER(roctracer_get_timestamp(×tamp)); return timestamp; } @@ -215,7 +210,7 @@ std::atomic_bool stop_flush_thread = false; void flush_thr_fun() { while (!stop_flush_thread) { - ROCTRACER_CALL(roctracer_flush_activity()); + CHECK_ROCTRACER(roctracer_flush_activity()); roctracer::TraceBufferBase::FlushAll(); std::this_thread::sleep_until(std::chrono::steady_clock::now() + std::chrono::microseconds(control_flush_us)); @@ -566,7 +561,7 @@ void pool_activity_callback(const char* begin, const char* end, void* arg) { } break; } - ROCTRACER_CALL(roctracer_next_record(record, &record)); + CHECK_ROCTRACER(roctracer_next_record(record, &record)); } } @@ -662,15 +657,15 @@ void open_tracing_pool() { roctracer_properties_t properties{}; properties.buffer_size = 0x80000; properties.buffer_callback_fun = pool_activity_callback; - ROCTRACER_CALL(roctracer_open_pool(&properties)); + CHECK_ROCTRACER(roctracer_open_pool(&properties)); } } // Flush tracing pool void close_tracing_pool() { if (roctracer_pool_t* pool = roctracer_default_pool(); pool != nullptr) { - ROCTRACER_CALL(roctracer_flush_activity_expl(pool)); - ROCTRACER_CALL(roctracer_close_pool_expl(pool)); + CHECK_ROCTRACER(roctracer_flush_activity_expl(pool)); + CHECK_ROCTRACER(roctracer_close_pool_expl(pool)); } } @@ -699,18 +694,18 @@ void tool_unload() { } if (trace_roctx) { - ROCTRACER_CALL(roctracer_disable_domain_callback(ACTIVITY_DOMAIN_ROCTX)); + CHECK_ROCTRACER(roctracer_disable_domain_callback(ACTIVITY_DOMAIN_ROCTX)); } if (trace_hsa_api) { - ROCTRACER_CALL(roctracer_disable_domain_callback(ACTIVITY_DOMAIN_HSA_API)); + CHECK_ROCTRACER(roctracer_disable_domain_callback(ACTIVITY_DOMAIN_HSA_API)); } if (trace_hsa_activity || trace_pcs) { - ROCTRACER_CALL(roctracer_disable_domain_activity(ACTIVITY_DOMAIN_HSA_OPS)); + CHECK_ROCTRACER(roctracer_disable_domain_activity(ACTIVITY_DOMAIN_HSA_OPS)); } if (trace_hip_api || trace_hip_activity) { - ROCTRACER_CALL(roctracer_disable_domain_callback(ACTIVITY_DOMAIN_HIP_API)); - ROCTRACER_CALL(roctracer_disable_domain_activity(ACTIVITY_DOMAIN_HIP_API)); - ROCTRACER_CALL(roctracer_disable_domain_activity(ACTIVITY_DOMAIN_HIP_OPS)); + CHECK_ROCTRACER(roctracer_disable_domain_callback(ACTIVITY_DOMAIN_HIP_API)); + CHECK_ROCTRACER(roctracer_disable_domain_activity(ACTIVITY_DOMAIN_HIP_API)); + CHECK_ROCTRACER(roctracer_disable_domain_activity(ACTIVITY_DOMAIN_HIP_OPS)); } // Flush tracing pool @@ -836,7 +831,7 @@ void tool_load() { // initialize HSA tracing fprintf(stdout, " rocTX-trace()\n"); fflush(stdout); - ROCTRACER_CALL( + CHECK_ROCTRACER( roctracer_enable_domain_callback(ACTIVITY_DOMAIN_ROCTX, roctx_api_callback, NULL)); } @@ -884,9 +879,9 @@ void tool_load() { } // HSA-runtime tool on-load method -extern "C" PUBLIC_API bool OnLoad(HsaApiTable* table, uint64_t runtime_version, - uint64_t failed_tool_count, - const char* const* failed_tool_names) { +extern "C" ROCTRACER_TOOL_EXPORT bool OnLoad(HsaApiTable* table, uint64_t runtime_version, + uint64_t failed_tool_count, + const char* const* failed_tool_names) { ONLOAD_TRACE_BEG(); const char* output_prefix = getenv("ROCP_OUTPUT_DIR"); @@ -932,13 +927,13 @@ extern "C" PUBLIC_API bool OnLoad(HsaApiTable* table, uint64_t runtime_version, for (unsigned i = 0; i < hsa_api_vec().size(); ++i) { uint32_t cid = HSA_API_ID_NUMBER; const char* api = hsa_api_vec()[i].c_str(); - ROCTRACER_CALL(roctracer_op_code(ACTIVITY_DOMAIN_HSA_API, api, &cid, NULL)); - ROCTRACER_CALL( + CHECK_ROCTRACER(roctracer_op_code(ACTIVITY_DOMAIN_HSA_API, api, &cid, NULL)); + CHECK_ROCTRACER( roctracer_enable_op_callback(ACTIVITY_DOMAIN_HSA_API, cid, hsa_api_callback, NULL)); printf(" %s", api); } } else { - ROCTRACER_CALL( + CHECK_ROCTRACER( roctracer_enable_domain_callback(ACTIVITY_DOMAIN_HSA_API, hsa_api_callback, NULL)); } printf(")\n"); @@ -958,7 +953,7 @@ extern "C" PUBLIC_API bool OnLoad(HsaApiTable* table, uint64_t runtime_version, fprintf(stdout, " HSA-activity-trace()\n"); fflush(stdout); - ROCTRACER_CALL(roctracer_enable_op_activity(ACTIVITY_DOMAIN_HSA_OPS, HSA_OP_ID_COPY)); + CHECK_ROCTRACER(roctracer_enable_op_activity(ACTIVITY_DOMAIN_HSA_OPS, HSA_OP_ID_COPY)); } // Enable HIP API callbacks/activity @@ -983,13 +978,13 @@ extern "C" PUBLIC_API bool OnLoad(HsaApiTable* table, uint64_t runtime_version, for (unsigned i = 0; i < hip_api_vec().size(); ++i) { uint32_t cid = HIP_API_ID_NONE; const char* api = hip_api_vec()[i].c_str(); - ROCTRACER_CALL(roctracer_op_code(ACTIVITY_DOMAIN_HIP_API, api, &cid, NULL)); - ROCTRACER_CALL( + CHECK_ROCTRACER(roctracer_op_code(ACTIVITY_DOMAIN_HIP_API, api, &cid, NULL)); + CHECK_ROCTRACER( roctracer_enable_op_callback(ACTIVITY_DOMAIN_HIP_API, cid, hip_api_callback, NULL)); printf(" %s", api); } } else { - ROCTRACER_CALL( + CHECK_ROCTRACER( roctracer_enable_domain_callback(ACTIVITY_DOMAIN_HIP_API, hip_api_callback, NULL)); } @@ -1006,7 +1001,7 @@ extern "C" PUBLIC_API bool OnLoad(HsaApiTable* table, uint64_t runtime_version, if (trace_hip_activity) { hcc_activity_file_handle = open_output_file(output_prefix, "hcc_ops_trace.txt"); - ROCTRACER_CALL(roctracer_enable_domain_activity(ACTIVITY_DOMAIN_HIP_OPS)); + CHECK_ROCTRACER(roctracer_enable_domain_activity(ACTIVITY_DOMAIN_HIP_OPS)); if (is_stats_opt) { FILE* f = NULL; @@ -1025,7 +1020,7 @@ extern "C" PUBLIC_API bool OnLoad(HsaApiTable* table, uint64_t runtime_version, fflush(stdout); open_tracing_pool(); pc_sample_file_handle = open_output_file(output_prefix, "pcs_trace.txt"); - ROCTRACER_CALL(roctracer_enable_op_activity(ACTIVITY_DOMAIN_HSA_OPS, HSA_OP_ID_RESERVED1)); + CHECK_ROCTRACER(roctracer_enable_op_activity(ACTIVITY_DOMAIN_HSA_OPS, HSA_OP_ID_RESERVED1)); } ONLOAD_TRACE_END(); @@ -1033,7 +1028,7 @@ extern "C" PUBLIC_API bool OnLoad(HsaApiTable* table, uint64_t runtime_version, } // HSA-runtime on-unload method -extern "C" PUBLIC_API void OnUnload() { ONLOAD_TRACE(""); } +extern "C" ROCTRACER_TOOL_EXPORT void OnUnload() { ONLOAD_TRACE(""); } extern "C" CONSTRUCTOR_API void constructor() { ONLOAD_TRACE_BEG(); diff --git a/projects/roctracer/test/app/MatrixTranspose_test.cpp b/projects/roctracer/test/app/MatrixTranspose_test.cpp index 840df63741..2d20ff2bdd 100644 --- a/projects/roctracer/test/app/MatrixTranspose_test.cpp +++ b/projects/roctracer/test/app/MatrixTranspose_test.cpp @@ -66,12 +66,23 @@ void SFLUSH() { // hip header file #include // Macro to call HIP API -#define HIP_CALL(call) \ +#define CALL_HIP(call) \ do { \ call; \ + } while (0); +#define CHECK_HIP(call) \ + do { \ + hipError_t err = call; \ + if (err != hipSuccess) { \ + fprintf(stderr, "%s\n", hipGetErrorString(err)); \ + abort(); \ + } \ } while (0) #else -#define HIP_CALL(call) \ +#define CALL_HIP(call) \ + do { \ + } while (0) +#define CHECK_HIP(call) \ do { \ } while (0) #endif @@ -142,7 +153,7 @@ int main() { hipSetDevice(devIndex); hipDeviceProp_t devProp; - HIP_CALL(hipGetDeviceProperties(&devProp, 0)); + CHECK_HIP(hipGetDeviceProperties(&devProp, 0)); fprintf(stderr, "Device %d name: %s\n", devIndex, devProp.name); #endif @@ -156,8 +167,8 @@ int main() { } // allocate the memory on the device side - HIP_CALL(hipMalloc((void**)&gpuMatrix, NUM * sizeof(float))); - HIP_CALL(hipMalloc((void**)&gpuTransposeMatrix, NUM * sizeof(float))); + CHECK_HIP(hipMalloc((void**)&gpuMatrix, NUM * sizeof(float))); + CHECK_HIP(hipMalloc((void**)&gpuTransposeMatrix, NUM * sizeof(float))); // correlation reagion32 roctracer_activity_push_external_correlation_id(31); @@ -165,7 +176,7 @@ int main() { roctracer_activity_push_external_correlation_id(32); // Memory transfer from host to device - HIP_CALL(hipMemcpy(gpuMatrix, Matrix, NUM * sizeof(float), hipMemcpyHostToDevice)); + CHECK_HIP(hipMemcpy(gpuMatrix, Matrix, NUM * sizeof(float), hipMemcpyHostToDevice)); // correlation reagion33 roctracer_activity_push_external_correlation_id(33); @@ -174,7 +185,7 @@ int main() { roctxRangePush("hipLaunchKernel"); // Lauching kernel from host - HIP_CALL(hipLaunchKernelGGL(matrixTranspose, + CALL_HIP(hipLaunchKernelGGL(matrixTranspose, dim3(WIDTH / THREADS_PER_BLOCK_X, WIDTH / THREADS_PER_BLOCK_Y), dim3(THREADS_PER_BLOCK_X, THREADS_PER_BLOCK_Y), 0, 0, gpuTransposeMatrix, gpuMatrix, WIDTH)); @@ -187,7 +198,7 @@ int main() { // Memory transfer from device to host roctxRangePush("hipMemcpy"); - HIP_CALL( + CHECK_HIP( hipMemcpy(TransposeMatrix, gpuTransposeMatrix, NUM * sizeof(float), hipMemcpyDeviceToHost)); roctxRangePop(); // for "hipMemcpy" @@ -197,7 +208,9 @@ int main() { roctracer_activity_pop_external_correlation_id(NULL); // CPU MatrixTranspose computation - HIP_CALL(matrixTransposeCPUReference(cpuTransposeMatrix, Matrix, WIDTH)); +#if HIP_TEST + matrixTransposeCPUReference(cpuTransposeMatrix, Matrix, WIDTH); +#endif // verify the results errors = 0; @@ -215,8 +228,8 @@ int main() { } // free the resources on device side - HIP_CALL(hipFree(gpuMatrix)); - HIP_CALL(hipFree(gpuTransposeMatrix)); + CHECK_HIP(hipFree(gpuMatrix)); + CHECK_HIP(hipFree(gpuTransposeMatrix)); // correlation reagion end roctracer_activity_pop_external_correlation_id(NULL); @@ -246,7 +259,7 @@ int main() { #include /* For SYS_xxx definitions */ // Macro to check ROC-tracer calls status -#define ROCTRACER_CALL(call) \ +#define CHECK_ROCTRACER(call) \ do { \ int err = call; \ if (err != 0) { \ @@ -333,7 +346,7 @@ void activity_callback(const char* begin, const char* end, void* arg) { SPRINT("\n"); SFLUSH(); - ROCTRACER_CALL(roctracer_next_record(record, &record)); + CHECK_ROCTRACER(roctracer_next_record(record, &record)); } } @@ -347,18 +360,18 @@ void init_tracing() { memset(&properties, 0, sizeof(roctracer_properties_t)); properties.buffer_size = 0x1000; properties.buffer_callback_fun = activity_callback; - ROCTRACER_CALL(roctracer_open_pool(&properties)); + CHECK_ROCTRACER(roctracer_open_pool(&properties)); // Enable HIP API callbacks - ROCTRACER_CALL(roctracer_enable_domain_callback(ACTIVITY_DOMAIN_HIP_API, api_callback, NULL)); + CHECK_ROCTRACER(roctracer_enable_domain_callback(ACTIVITY_DOMAIN_HIP_API, api_callback, NULL)); // Enable HIP activity tracing #if HIP_API_ACTIVITY_ON - ROCTRACER_CALL(roctracer_enable_domain_activity(ACTIVITY_DOMAIN_HIP_API)); + CHECK_ROCTRACER(roctracer_enable_domain_activity(ACTIVITY_DOMAIN_HIP_API)); #endif - ROCTRACER_CALL(roctracer_enable_domain_activity(ACTIVITY_DOMAIN_HIP_OPS)); + CHECK_ROCTRACER(roctracer_enable_domain_activity(ACTIVITY_DOMAIN_HIP_OPS)); // Enable PC sampling - ROCTRACER_CALL(roctracer_enable_op_activity(ACTIVITY_DOMAIN_HSA_OPS, HSA_OP_ID_RESERVED1)); + CHECK_ROCTRACER(roctracer_enable_op_activity(ACTIVITY_DOMAIN_HSA_OPS, HSA_OP_ID_RESERVED1)); // Enable rocTX - ROCTRACER_CALL(roctracer_enable_domain_callback(ACTIVITY_DOMAIN_ROCTX, api_callback, NULL)); + CHECK_ROCTRACER(roctracer_enable_domain_callback(ACTIVITY_DOMAIN_ROCTX, api_callback, NULL)); } // Start tracing routine @@ -373,14 +386,14 @@ void start_tracing() { // Stop tracing routine void stop_tracing() { - ROCTRACER_CALL(roctracer_disable_domain_callback(ACTIVITY_DOMAIN_HIP_API)); + CHECK_ROCTRACER(roctracer_disable_domain_callback(ACTIVITY_DOMAIN_HIP_API)); #if HIP_API_ACTIVITY_ON - ROCTRACER_CALL(roctracer_disable_domain_activity(ACTIVITY_DOMAIN_HIP_API)); + CHECK_ROCTRACER(roctracer_disable_domain_activity(ACTIVITY_DOMAIN_HIP_API)); #endif - ROCTRACER_CALL(roctracer_disable_domain_activity(ACTIVITY_DOMAIN_HIP_OPS)); - ROCTRACER_CALL(roctracer_disable_domain_activity(ACTIVITY_DOMAIN_HSA_OPS)); - ROCTRACER_CALL(roctracer_disable_domain_callback(ACTIVITY_DOMAIN_ROCTX)); - ROCTRACER_CALL(roctracer_flush_activity()); + CHECK_ROCTRACER(roctracer_disable_domain_activity(ACTIVITY_DOMAIN_HIP_OPS)); + CHECK_ROCTRACER(roctracer_disable_domain_activity(ACTIVITY_DOMAIN_HSA_OPS)); + CHECK_ROCTRACER(roctracer_disable_domain_callback(ACTIVITY_DOMAIN_ROCTX)); + CHECK_ROCTRACER(roctracer_flush_activity()); fprintf(stderr, "# STOP #############################\n"); } #else