Skip to content

Commit cc8784d

Browse files
wooway777wooway777
committed
issue/571 - low-level event related interfaces
1 parent 2d0a83c commit cc8784d

File tree

14 files changed

+203
-5
lines changed

14 files changed

+203
-5
lines changed

‎include/infinicore/context/context.hpp‎

Lines changed: 9 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -30,6 +30,15 @@ void memcpyD2H(void *dst, const void *src, size_t size);
3030
void memcpyD2D(void *dst, const void *src, size_t size);
3131
void memcpyH2H(void *dst, const void *src, size_t size);
3232

33+
// Timing APIs for performance measurement
34+
infinirtEvent_t createEvent();
35+
infinirtEvent_t createEventWithFlags(uint32_t flags);
36+
void recordEvent(infinirtEvent_t event, infinirtStream_t stream = nullptr);
37+
bool queryEvent(infinirtEvent_t event);
38+
void synchronizeEvent(infinirtEvent_t event);
39+
void destroyEvent(infinirtEvent_t event);
40+
float elapsedTime(infinirtEvent_t start, infinirtEvent_t end);
41+
3342
} // namespace context
3443

3544
} // namespace infinicore

‎include/infinirt.h‎

Lines changed: 10 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -2,6 +2,7 @@
22
#define __INFINIRT_API_H__
33

44
#include "infinicore.h"
5+
#include <stdint.h>
56

67
typedef void *infinirtStream_t;
78
typedef void *infinirtEvent_t;
@@ -27,11 +28,20 @@ typedef enum {
2728
INFINIRT_EVENT_NOT_READY = 1,
2829
} infinirtEventStatus_t;
2930

31+
// Event flags for precise timing
32+
typedef enum {
33+
INFINIRT_EVENT_DEFAULT = 0x0, // Default event creation flags
34+
INFINIRT_EVENT_DISABLE_TIMING = 0x1, // Event will not record timing data
35+
INFINIRT_EVENT_BLOCKING_SYNC = 0x2, // Event uses blocking synchronization
36+
} infinirtEventFlags_t;
37+
3038
__C __export infiniStatus_t infinirtEventCreate(infinirtEvent_t *event_ptr);
39+
__C __export infiniStatus_t infinirtEventCreateWithFlags(infinirtEvent_t *event_ptr, uint32_t flags);
3140
__C __export infiniStatus_t infinirtEventRecord(infinirtEvent_t event, infinirtStream_t stream);
3241
__C __export infiniStatus_t infinirtEventQuery(infinirtEvent_t event, infinirtEventStatus_t *status_ptr);
3342
__C __export infiniStatus_t infinirtEventSynchronize(infinirtEvent_t event);
3443
__C __export infiniStatus_t infinirtEventDestroy(infinirtEvent_t event);
44+
__C __export infiniStatus_t infinirtEventElapsedTime(float *ms_ptr, infinirtEvent_t start, infinirtEvent_t end);
3545

3646
// Memory
3747
typedef enum {

‎src/infinicore/context/context_impl.cc‎

Lines changed: 29 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -139,6 +139,35 @@ void memcpyH2H(void *dst, const void *src, size_t size) {
139139
return ContextImpl::singleton().getCpuRuntime()->memcpyD2D(dst, src, size);
140140
}
141141

142+
// Timing API implementations
143+
infinirtEvent_t createEvent() {
144+
return ContextImpl::singleton().getCurrentRuntime()->createEvent();
145+
}
146+
147+
infinirtEvent_t createEventWithFlags(uint32_t flags) {
148+
return ContextImpl::singleton().getCurrentRuntime()->createEventWithFlags(flags);
149+
}
150+
151+
void recordEvent(infinirtEvent_t event, infinirtStream_t stream) {
152+
ContextImpl::singleton().getCurrentRuntime()->recordEvent(event, stream);
153+
}
154+
155+
bool queryEvent(infinirtEvent_t event) {
156+
return ContextImpl::singleton().getCurrentRuntime()->queryEvent(event);
157+
}
158+
159+
void synchronizeEvent(infinirtEvent_t event) {
160+
ContextImpl::singleton().getCurrentRuntime()->synchronizeEvent(event);
161+
}
162+
163+
void destroyEvent(infinirtEvent_t event) {
164+
ContextImpl::singleton().getCurrentRuntime()->destroyEvent(event);
165+
}
166+
167+
float elapsedTime(infinirtEvent_t start, infinirtEvent_t end) {
168+
return ContextImpl::singleton().getCurrentRuntime()->elapsedTime(start, end);
169+
}
170+
142171
} // namespace context
143172

144173
} // namespace infinicore

‎src/infinicore/context/runtime/runtime.cc‎

Lines changed: 40 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -88,6 +88,46 @@ void Runtime::memcpyD2D(void *dst, const void *src, size_t size) {
8888
INFINICORE_CHECK_ERROR(infinirtMemcpyAsync(dst, src, size, INFINIRT_MEMCPY_D2D, stream_));
8989
}
9090

91+
// Timing method implementations
92+
infinirtEvent_t Runtime::createEvent() {
93+
infinirtEvent_t event;
94+
INFINICORE_CHECK_ERROR(infinirtEventCreate(&event));
95+
return event;
96+
}
97+
98+
infinirtEvent_t Runtime::createEventWithFlags(uint32_t flags) {
99+
infinirtEvent_t event;
100+
INFINICORE_CHECK_ERROR(infinirtEventCreateWithFlags(&event, flags));
101+
return event;
102+
}
103+
104+
void Runtime::recordEvent(infinirtEvent_t event, infinirtStream_t stream) {
105+
if (stream == nullptr) {
106+
stream = stream_;
107+
}
108+
INFINICORE_CHECK_ERROR(infinirtEventRecord(event, stream));
109+
}
110+
111+
bool Runtime::queryEvent(infinirtEvent_t event) {
112+
infinirtEventStatus_t status;
113+
INFINICORE_CHECK_ERROR(infinirtEventQuery(event, &status));
114+
return status == INFINIRT_EVENT_COMPLETE;
115+
}
116+
117+
void Runtime::synchronizeEvent(infinirtEvent_t event) {
118+
INFINICORE_CHECK_ERROR(infinirtEventSynchronize(event));
119+
}
120+
121+
void Runtime::destroyEvent(infinirtEvent_t event) {
122+
INFINICORE_CHECK_ERROR(infinirtEventDestroy(event));
123+
}
124+
125+
float Runtime::elapsedTime(infinirtEvent_t start, infinirtEvent_t end) {
126+
float ms;
127+
INFINICORE_CHECK_ERROR(infinirtEventElapsedTime(&ms, start, end));
128+
return ms;
129+
}
130+
91131
std::string Runtime::toString() const {
92132
return fmt::format("Runtime({})", device_.toString());
93133
}

‎src/infinicore/context/runtime/runtime.hpp‎

Lines changed: 9 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -38,6 +38,15 @@ class Runtime {
3838
void memcpyD2H(void *dst, const void *src, size_t size);
3939
void memcpyD2D(void *dst, const void *src, size_t size);
4040

41+
// Timing methods
42+
infinirtEvent_t createEvent();
43+
infinirtEvent_t createEventWithFlags(uint32_t flags);
44+
void recordEvent(infinirtEvent_t event, infinirtStream_t stream = nullptr);
45+
bool queryEvent(infinirtEvent_t event);
46+
void synchronizeEvent(infinirtEvent_t event);
47+
void destroyEvent(infinirtEvent_t event);
48+
float elapsedTime(infinirtEvent_t start, infinirtEvent_t end);
49+
4150
std::string toString() const;
4251

4352
friend class ContextImpl;

‎src/infinirt/ascend/infinirt_ascend.cc‎

Lines changed: 8 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -64,6 +64,10 @@ infiniStatus_t eventCreate(infinirtEvent_t *event_ptr) {
6464
return INFINI_STATUS_SUCCESS;
6565
}
6666

67+
infiniStatus_t eventCreateWithFlags(infinirtEvent_t *event_ptr, uint32_t flags) {
68+
return INFINI_STATUS_NOT_IMPLEMENTED;
69+
}
70+
6771
infiniStatus_t eventRecord(infinirtEvent_t event, infinirtStream_t stream) {
6872
CHECK_ACLRT(aclrtRecordEvent((aclrtEvent)event, (aclrtStream)stream));
6973
return INFINI_STATUS_SUCCESS;
@@ -90,6 +94,10 @@ infiniStatus_t eventDestroy(infinirtEvent_t event) {
9094
return INFINI_STATUS_SUCCESS;
9195
}
9296

97+
infiniStatus_t eventElapsedTime(float *ms_ptr, infinirtEvent_t start, infinirtEvent_t end) {
98+
return INFINI_STATUS_NOT_IMPLEMENTED;
99+
}
100+
93101
infiniStatus_t mallocDevice(void **p_ptr, size_t size) {
94102
CHECK_ACLRT(aclrtMallocAlign32(p_ptr, size, ACL_MEM_MALLOC_HUGE_FIRST));
95103
return INFINI_STATUS_SUCCESS;

‎src/infinirt/bang/infinirt_bang.cc‎

Lines changed: 8 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -51,6 +51,10 @@ infiniStatus_t eventCreate(infinirtEvent_t *event_ptr) {
5151
return INFINI_STATUS_SUCCESS;
5252
}
5353

54+
infiniStatus_t eventCreateWithFlags(infinirtEvent_t *event_ptr, uint32_t flags) {
55+
return INFINI_STATUS_NOT_IMPLEMENTED;
56+
}
57+
5458
infiniStatus_t eventRecord(infinirtEvent_t event, infinirtStream_t stream) {
5559
CHECK_BANGRT(cnrtPlaceNotifier((cnrtNotifier_t)event, (cnrtQueue_t)stream));
5660
return INFINI_STATUS_SUCCESS;
@@ -78,6 +82,10 @@ infiniStatus_t eventDestroy(infinirtEvent_t event) {
7882
return INFINI_STATUS_SUCCESS;
7983
}
8084

85+
infiniStatus_t eventElapsedTime(float *ms_ptr, infinirtEvent_t start, infinirtEvent_t end) {
86+
return INFINI_STATUS_NOT_IMPLEMENTED;
87+
}
88+
8189
infiniStatus_t mallocDevice(void **p_ptr, size_t size) {
8290
CHECK_BANGRT(cnrtMalloc(p_ptr, size));
8391
return INFINI_STATUS_SUCCESS;

‎src/infinirt/cpu/infinirt_cpu.cc‎

Lines changed: 33 additions & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,5 @@
11
#include "infinirt_cpu.h"
2+
#include <chrono>
23
#include <cstdlib>
34
#include <cstring>
45

@@ -34,23 +35,50 @@ infiniStatus_t streamWaitEvent(infinirtStream_t stream, infinirtEvent_t event) {
3435
}
3536

3637
infiniStatus_t eventCreate(infinirtEvent_t *event_ptr) {
37-
return INFINI_STATUS_NOT_IMPLEMENTED;
38+
// For CPU implementation, we use a simple timestamp as event
39+
auto now = std::chrono::steady_clock::now();
40+
auto *timestamp = new std::chrono::steady_clock::time_point(now);
41+
*event_ptr = timestamp;
42+
return INFINI_STATUS_SUCCESS;
43+
}
44+
45+
infiniStatus_t eventCreateWithFlags(infinirtEvent_t *event_ptr, uint32_t flags) {
46+
// CPU implementation ignores flags for simplicity
47+
return eventCreate(event_ptr);
3848
}
3949

4050
infiniStatus_t eventRecord(infinirtEvent_t event, infinirtStream_t stream) {
41-
return INFINI_STATUS_NOT_IMPLEMENTED;
51+
// Update the event timestamp
52+
auto *timestamp = static_cast<std::chrono::steady_clock::time_point *>(event);
53+
*timestamp = std::chrono::steady_clock::now();
54+
return INFINI_STATUS_SUCCESS;
4255
}
4356

4457
infiniStatus_t eventQuery(infinirtEvent_t event, infinirtEventStatus_t *status_ptr) {
45-
return INFINI_STATUS_NOT_IMPLEMENTED;
58+
// CPU events are always complete immediately
59+
*status_ptr = INFINIRT_EVENT_COMPLETE;
60+
return INFINI_STATUS_SUCCESS;
4661
}
4762

4863
infiniStatus_t eventSynchronize(infinirtEvent_t event) {
49-
return INFINI_STATUS_NOT_IMPLEMENTED;
64+
// CPU events are synchronized immediately
65+
return INFINI_STATUS_SUCCESS;
5066
}
5167

5268
infiniStatus_t eventDestroy(infinirtEvent_t event) {
53-
return INFINI_STATUS_NOT_IMPLEMENTED;
69+
auto *timestamp = static_cast<std::chrono::steady_clock::time_point *>(event);
70+
delete timestamp;
71+
return INFINI_STATUS_SUCCESS;
72+
}
73+
74+
infiniStatus_t eventElapsedTime(float *ms_ptr, infinirtEvent_t start, infinirtEvent_t end) {
75+
auto *start_time = static_cast<std::chrono::steady_clock::time_point *>(start);
76+
auto *end_time = static_cast<std::chrono::steady_clock::time_point *>(end);
77+
78+
auto duration = std::chrono::duration_cast<std::chrono::microseconds>(*end_time - *start_time);
79+
*ms_ptr = static_cast<float>(duration.count()) / 1000.0f; // Convert microseconds to milliseconds
80+
81+
return INFINI_STATUS_SUCCESS;
5482
}
5583

5684
infiniStatus_t mallocDevice(void **p_ptr, size_t size) {

‎src/infinirt/cuda/infinirt_cuda.cu‎

Lines changed: 22 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -53,6 +53,23 @@ infiniStatus_t eventCreate(infinirtEvent_t *event_ptr) {
5353
return INFINI_STATUS_SUCCESS;
5454
}
5555

56+
infiniStatus_t eventCreateWithFlags(infinirtEvent_t *event_ptr, uint32_t flags) {
57+
cudaEvent_t event;
58+
unsigned int cuda_flags = cudaEventDefault;
59+
60+
// Convert infinirt flags to CUDA flags
61+
if (flags & INFINIRT_EVENT_DISABLE_TIMING) {
62+
cuda_flags |= cudaEventDisableTiming;
63+
}
64+
if (flags & INFINIRT_EVENT_BLOCKING_SYNC) {
65+
cuda_flags |= cudaEventBlockingSync;
66+
}
67+
68+
CHECK_CUDART(cudaEventCreateWithFlags(&event, cuda_flags));
69+
*event_ptr = event;
70+
return INFINI_STATUS_SUCCESS;
71+
}
72+
5673
infiniStatus_t eventRecord(infinirtEvent_t event, infinirtStream_t stream) {
5774
CHECK_CUDART(cudaEventRecord((cudaEvent_t)event, (cudaStream_t)stream));
5875
return INFINI_STATUS_SUCCESS;
@@ -80,6 +97,11 @@ infiniStatus_t eventDestroy(infinirtEvent_t event) {
8097
return INFINI_STATUS_SUCCESS;
8198
}
8299

100+
infiniStatus_t eventElapsedTime(float *ms_ptr, infinirtEvent_t start, infinirtEvent_t end) {
101+
CHECK_CUDART(cudaEventElapsedTime(ms_ptr, (cudaEvent_t)start, (cudaEvent_t)end));
102+
return INFINI_STATUS_SUCCESS;
103+
}
104+
83105
infiniStatus_t mallocDevice(void **p_ptr, size_t size) {
84106
CHECK_CUDART(cudaMalloc(p_ptr, size));
85107
return INFINI_STATUS_SUCCESS;

‎src/infinirt/infinirt.cc‎

Lines changed: 8 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -126,6 +126,10 @@ __C infiniStatus_t infinirtEventCreate(infinirtEvent_t *event_ptr) {
126126
INFINIRT_CALL_DEVICE_API(eventCreate, (event_ptr));
127127
}
128128

129+
__C infiniStatus_t infinirtEventCreateWithFlags(infinirtEvent_t *event_ptr, uint32_t flags) {
130+
INFINIRT_CALL_DEVICE_API(eventCreateWithFlags, (event_ptr, flags));
131+
}
132+
129133
__C infiniStatus_t infinirtEventRecord(infinirtEvent_t event, infinirtStream_t stream) {
130134
INFINIRT_CALL_DEVICE_API(eventRecord, (event, stream));
131135
}
@@ -142,6 +146,10 @@ __C infiniStatus_t infinirtEventDestroy(infinirtEvent_t event) {
142146
INFINIRT_CALL_DEVICE_API(eventDestroy, (event));
143147
}
144148

149+
__C infiniStatus_t infinirtEventElapsedTime(float *ms_ptr, infinirtEvent_t start, infinirtEvent_t end) {
150+
INFINIRT_CALL_DEVICE_API(eventElapsedTime, (ms_ptr, start, end));
151+
}
152+
145153
__C infiniStatus_t infinirtMalloc(void **p_ptr, size_t size) {
146154
INFINIRT_CALL_DEVICE_API(mallocDevice, (p_ptr, size));
147155
}

0 commit comments

Comments
 (0)