AI/rocm-systems
2
0
Forkuj 0
Kod Zgłoszenia Oczekujące zmiany Actions Packages Projekty Wydania Wiki Aktywność

Add ROCtracer version information

Przeglądaj źródła 
Change-Id: I10b268790d2dc4f3a3ad8624b2f553da6f3ccc8e


[ROCm/roctracer commit: 1c450082af]
This commit is contained in:
Laurent Morichetti
2022-05-23 20:38:54 -07:00
zatwierdzone przez Laurent Morichetti
rodzic 23893311af
commit ef717b9a2a
6 zmienionych plików z 297 dodań i 257 usunięć
Ściągnij plik aktualizacji Ściągnij plik porównania Expand all files Collapse all files
projects/roctracer/inc/ext/prof_protocol.h
+38 -37
Wyświetl plik
@@ -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 <stdlib.h>
// 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_ */
projects/roctracer/inc/roctracer.h
+120 -87
Wyświetl plik
@@ -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 <stdint.h>
#include <stddef.h>
#ifndef __cplusplus
#include <stdbool.h>
#endif
#include <ext/prof_protocol.h>
#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_ */
projects/roctracer/inc/roctracer_ext.h
+9 -8
Wyświetl plik
@@ -27,8 +27,8 @@
//
////////////////////////////////////////////////////////////////////////////////
#ifndef INC_ROCTRACER_EXT_H_
#define INC_ROCTRACER_EXT_H_
#ifndef ROCTRACER_EXT_H_
#define ROCTRACER_EXT_H_
#include <roctracer.h>
@@ -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_
projects/roctracer/src/roctracer/roctracer.cpp
+58 -61
Wyświetl plik
@@ -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<roctracer_pool_t*>(default_memory_pool);
if (pool != nullptr) default_memory_pool = reinterpret_cast<MemoryPool*>(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<roctracer_pool_t*>(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() {}
extern "C" ROCTRACER_EXPORT void OnUnload() {}
projects/roctracer/src/tracer_tool/tracer_tool.cpp
+34 -39
Wyświetl plik
@@ -27,7 +27,6 @@
#include <cxxabi.h> /* names denangle */
#include <dirent.h>
#include <pthread.h>
#include <stdarg.h>
#include <stdio.h>
#include <string.h>
@@ -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(&timestamp));
CHECK_ROCTRACER(roctracer_get_timestamp(&timestamp));
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();
projects/roctracer/test/app/MatrixTranspose_test.cpp
+38 -25
Wyświetl plik
@@ -66,12 +66,23 @@ void SFLUSH() {
// hip header file
#include <hip/hip_runtime.h>
// 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 <sys/syscall.h> /* 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