-
Notifications
You must be signed in to change notification settings - Fork 205
fix: implement short-term fixes for lock contention and add tests #285
New issue
Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.
By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.
Already on GitHub? Sign in to your account
base: main
Are you sure you want to change the base?
Changes from all commits
60792aa
f58fac0
f0a6515
8803f60
f49ad8e
ad9f4cb
6e3ae85
1947b3a
31f3c11
7fe0387
File filter
Filter by extension
Conversations
Jump to
Diff view
Diff view
There are no files selected for viewing
| Original file line number | Diff line number | Diff line change |
|---|---|---|
|
|
@@ -36,6 +36,12 @@ foreach(TEST_SCRIPT ${TEST_SCRIPTS}) | |
| list(APPEND TEST_TARGET_NAMES_LIST ${TEST_TARGET_NAME}) | ||
| if (TEST_TARGET_NAME STREQUAL "test_postinit_owner_death") | ||
| target_link_libraries(${TEST_TARGET_NAME} -lrt -lpthread) | ||
| elseif(TEST_TARGET_NAME STREQUAL "bench_lock") | ||
|
Contributor
There was a problem hiding this comment. Choose a reason for hiding this commentThe reason will be displayed to describe this comment to others. Learn more. bench_lock.c is not in this pr. dead block. add the file or remove this.
Author
There was a problem hiding this comment. Choose a reason for hiding this commentThe reason will be displayed to describe this comment to others. Learn more. yes sorry it was in a previous commit i forgot to remove it |
||
| # Tell CMake to compile utils.c together with bench_lock.c | ||
| target_sources(${TEST_TARGET_NAME} PRIVATE ${CMAKE_CURRENT_SOURCE_DIR}/../src/utils.c) | ||
| # Add the include directory so it can find the headers | ||
| target_include_directories(${TEST_TARGET_NAME} PRIVATE ${CMAKE_CURRENT_SOURCE_DIR}/../src) | ||
| target_link_libraries(${TEST_TARGET_NAME} -lrt -lpthread) | ||
| else() | ||
| target_link_libraries(${TEST_TARGET_NAME} -lrt -lpthread | ||
| -lnvidia-ml -lcuda -lcudart -L${CUDA_HOME}/lib64) | ||
|
|
||
| Original file line number | Diff line number | Diff line change | ||||||||||||||||||||||||||||||||||||||||||||||||||
|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|---|
| @@ -0,0 +1,45 @@ | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| #include <stdio.h> | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| #include <stdlib.h> | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| #include <unistd.h> | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| #include <sys/wait.h> | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| #include <cuda_runtime.h> | ||||||||||||||||||||||||||||||||||||||||||||||||||||
|
|
||||||||||||||||||||||||||||||||||||||||||||||||||||
| #define NUM_WORKERS 4 | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| #define ITERATIONS 10 | ||||||||||||||||||||||||||||||||||||||||||||||||||||
|
|
||||||||||||||||||||||||||||||||||||||||||||||||||||
| void worker_task(int worker_id) { | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| void *ptr = NULL; | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| for (int i = 0; i < ITERATIONS; i++) { | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| // Allocate 1MB to trigger HAMi-core's memory interception | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| cudaError_t err = cudaMalloc(&ptr, 1024 * 1024); | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| if (err != cudaSuccess) { | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| printf("Worker %d failed allocation: %s\n", worker_id, cudaGetErrorString(err)); | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| exit(1); | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| } | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| cudaFree(ptr); | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| } | ||||||||||||||||||||||||||||||||||||||||||||||||||||
|
Comment on lines
+12
to
+20
Contributor
There was a problem hiding this comment. Choose a reason for hiding this commentThe reason will be displayed to describe this comment to others. Learn more. 🎯 Functional Correctness | 🟠 Major | 🏗️ Heavy lift 🧩 Analysis chain🏁 Script executed: #!/bin/bash
set -euo pipefail
# Locate the concurrent test registration and memory-limit configuration.
fd -a 'CMakeLists.txt|Makefile' . -x rg -n -C 4 \
'test_alloc_concurrent|cudaErrorMemoryAllocation|memory.*limit|vgpu|HAMi' {}Repository: Project-HAMi/HAMi-core Length of output: 2099 🏁 Script executed: #!/bin/bash
set -euo pipefail
printf '%s\n' '--- candidate files ---'
fd -a -t f 'test_alloc_concurrent\.cu|CMakeLists\.txt|.*memory.*limit.*|.*alloc.*concurrent.*' .
printf '%s\n' '--- test source ---'
test_file="$(fd -a -t f 'test_alloc_concurrent\.cu' | head -n 1)"
cat -n "$test_file"
printf '%s\n' '--- related registrations and limit configuration ---'
rg -n -C 6 \
'test_alloc_concurrent|cudaErrorMemoryAllocation|memory.*limit|CUDA_VISIBLE_DEVICES|vgpu|HAMi|set.*limit|limit.*set' \
--glob 'CMakeLists.txt' --glob '*.cmake' --glob '*.cu' --glob '*.cpp' --glob '*.c' --glob '*.h' .Repository: Project-HAMi/HAMi-core Length of output: 50378 🏁 Script executed: #!/bin/bash
set -euo pipefail
printf '%s\n' '--- test/CMakeLists.txt ---'
cat -n test/CMakeLists.txt
printf '%s\n' '--- memory-limit environment and setters ---'
rg -n -C 5 \
'CUDA_DEVICE_MEMORY_LIMIT|MEMORY_LIMIT|set_current_device_memory_limit|get_current_device_memory_limit|oom_check' \
src test README.md Makefile CMakeLists.txt 2>/dev/null | head -n 240
printf '%s\n' '--- allocation interception ---'
rg -n -C 8 \
'cudaMalloc|cuMemAlloc|allocate_raw|oom_check' \
src/cuda src/allocator test/test_alloc_concurrent.cu | head -n 260
printf '%s\n' '--- concurrent-test references ---'
rg -n -C 8 'test_alloc_concurrent|alloc_concurrent' . --glob '!*.cu' --glob '!*.c' --glob '!*.h' --glob '!*.md'Repository: Project-HAMi/HAMi-core Length of output: 33048 🏁 Script executed: #!/bin/bash
set -euo pipefail
printf '%s\n' '--- limit initialization and cache behavior ---'
sed -n '120,215p' src/multiprocess/multiprocess_memory_limit.c
sed -n '1,80p' src/multiprocess/multiprocess_memory_limit.c
rg -n -C 6 \
'do_init_device_memory_limits|CUDA_DEVICE_MEMORY_LIMIT|shared.*cache|CACHE_DEFAULT|create.*region|open.*region' \
src/multiprocess/multiprocess_memory_limit.c src/multiprocess/multiprocess_memory_limit.h
printf '%s\n' '--- relevant driver/runtime hooks ---'
rg -n -C 8 \
'cuMemAlloc_v2|cudaMalloc|CUDA_ERROR_OUT_OF_MEMORY|cudaErrorMemoryAllocation|add_chunk\(' \
src test --glob '*.[ch]' --glob '*.cu' | head -n 220
printf '%s\n' '--- static behavioral verifier ---'
python3 - <<'PY'
from pathlib import Path
import re
p = next(Path('.').rglob('test_alloc_concurrent.cu'))
s = p.read_text()
loop = re.search(
r'for\s*\([^)]*i\s*<\s*ITERATIONS[^)]*\)\s*\{(?P<body>.*?)\n\s*\}',
s, re.S,
)
assert loop, "worker iteration loop not found"
body = loop.group('body')
assert re.search(r'cudaMalloc\s*\(', body), "loop does not allocate"
assert re.search(r'cudaFree\s*\(\s*ptr\s*\)', body), "loop does not free ptr"
assert 'cudaErrorMemoryAllocation' not in s, "OOM expectation already exists"
wait_loop = re.search(r'while\s*\(\s*wait\s*\(\s*&status\s*\)\s*>\s*0\s*\)\s*;', s)
assert wait_loop, "parent wait loop not found"
parent_tail = s[wait_loop.end():]
assert 'WIFEXITED' not in parent_tail and 'WEXITSTATUS' not in parent_tail
assert re.search(r'return\s+0\s*;', parent_tail), "parent success return not found"
print("allocation is freed within each iteration")
print("no cudaErrorMemoryAllocation expectation exists")
print("parent waits but does not inspect child status before returning success")
PYRepository: Project-HAMi/HAMi-core Length of output: 28841 🏁 Script executed: #!/bin/bash
set -euo pipefail
python3 - <<'PY'
from pathlib import Path
import re
p = next(Path('.').rglob('test_alloc_concurrent.cu'))
s = p.read_text()
start = s.index('void worker_task')
end = s.index('\nint main', start)
worker = s[start:end]
assert 'cudaMalloc(&ptr, 1024 * 1024)' in worker
assert 'cudaFree(ptr);' in worker
assert worker.index('cudaMalloc') < worker.index('cudaFree')
assert 'cudaErrorMemoryAllocation' not in s
main = s[s.index('int main'): ]
wait = 'while (wait(&status) > 0);'
assert wait in main
after_wait = main[main.index(wait) + len(wait):]
assert 'WIFEXITED' not in after_wait
assert 'WEXITSTATUS' not in after_wait
assert re.search(r'\breturn\s+0\s*;', after_wait)
print('worker frees each successful allocation before the next iteration')
print('test has no expected cudaErrorMemoryAllocation assertion')
print('parent ignores child status and returns success after wait')
cmake = Path('test/CMakeLists.txt').read_text()
registered = re.findall(r'add_test\s*\(\s*NAME\s+([A-Za-z0-9_.-]+)', cmake)
print('CTest registrations:', registered)
assert 'test_alloc_concurrent' not in registered
assert 'postinit_owner_death' in registered
print('test_alloc_concurrent is compiled by the source glob but is not registered with CTest')
PY
printf '%s\n' '--- exact CTest registration section ---'
sed -n '7,60p' test/CMakeLists.txtRepository: Project-HAMi/HAMi-core Length of output: 3228 Make
🤖 Prompt for AI Agents |
||||||||||||||||||||||||||||||||||||||||||||||||||||
| printf("Worker %d completed successfully.\n", worker_id); | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| exit(0); | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| } | ||||||||||||||||||||||||||||||||||||||||||||||||||||
|
|
||||||||||||||||||||||||||||||||||||||||||||||||||||
| int main() { | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| printf("Starting %d concurrent workers...\n", NUM_WORKERS); | ||||||||||||||||||||||||||||||||||||||||||||||||||||
|
|
||||||||||||||||||||||||||||||||||||||||||||||||||||
|
Check failure on line 27 in test/test_alloc_concurrent.cu
|
||||||||||||||||||||||||||||||||||||||||||||||||||||
| for (int i = 0; i < NUM_WORKERS; i++) { | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| pid_t pid = fork(); | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| if (pid == 0) { | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| // Child process | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| worker_task(i); | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| } else if (pid < 0) { | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| printf("Fork failed.\n"); | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| return 1; | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| } | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| } | ||||||||||||||||||||||||||||||||||||||||||||||||||||
|
|
||||||||||||||||||||||||||||||||||||||||||||||||||||
| // Parent waits for all children to finish | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| int status; | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| while (wait(&status) > 0); | ||||||||||||||||||||||||||||||||||||||||||||||||||||
|
Check failure on line 41 in test/test_alloc_concurrent.cu
|
||||||||||||||||||||||||||||||||||||||||||||||||||||
|
|
||||||||||||||||||||||||||||||||||||||||||||||||||||
|
Check failure on line 42 in test/test_alloc_concurrent.cu
|
||||||||||||||||||||||||||||||||||||||||||||||||||||
| printf("All workers finished.\n"); | ||||||||||||||||||||||||||||||||||||||||||||||||||||
| return 0; | ||||||||||||||||||||||||||||||||||||||||||||||||||||
|
Comment on lines
+39
to
+44
Contributor
There was a problem hiding this comment. Choose a reason for hiding this commentThe reason will be displayed to describe this comment to others. Learn more. 🎯 Functional Correctness | 🟠 Major | ⚡ Quick win Propagate worker failures to the test result. A child that calls Proposed fix- while (wait(&status) > 0);
+ int failed = 0;
+ for (int completed = 0; completed < NUM_WORKERS; ) {
+ if (wait(&status) < 0) {
+ perror("wait");
+ return 1;
+ }
+ completed++;
+ if (!WIFEXITED(status) || WEXITSTATUS(status) != 0) {
+ failed = 1;
+ }
+ }
+
+ if (failed) {
+ return 1;
+ }📝 Committable suggestion
Suggested change
🤖 Prompt for AI Agents |
||||||||||||||||||||||||||||||||||||||||||||||||||||
| } | ||||||||||||||||||||||||||||||||||||||||||||||||||||
|
Check failure on line 45 in test/test_alloc_concurrent.cu
|
||||||||||||||||||||||||||||||||||||||||||||||||||||
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
🎯 Functional Correctness | 🔴 Critical | ⚡ Quick win
Update all lock descriptor references.
Line 19 declares
unified_lock_fd, but the lock and unlock functions still uselock_fd. This causes an undeclared-identifier build failure. Update all references tounified_lock_fd, or retain the original declaration name. Thebench_locktarget compiles this file directly.🤖 Prompt for AI Agents