augustin64/tipe
1
1
Fork 0
mirror of https://github.com/augustin64/projet-tipe synced 2025-01-23 15:16:26 +01:00
Code Issues Packages Projects Releases Wiki Activity

Add cuda_memory_management test

Browse Source
This commit is contained in:
augustin64 2023-02-19 10:31:19 +01:00
parent 2d46136609
commit 6a98d2d9cf
3 changed files with 110 additions and 4 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

16
Makefile
View File

@ -161,26 +161,34 @@ run-tests: build-tests
$(foreach file, $(wildcard $(TEST_SRCDIR)/*.sh), $(file);)
@echo "$$(for file in build/test-*; do echo -e \\033[33m#####\\033[0m $$file \\033[33m#####\\033[0m; $$file || echo -e "\\033[1m\\033[31mErreur sur $$file\\033[0m"; done)"
build-tests: prepare-tests $(TESTS_OBJ) $(BUILDDIR)/test-cnn_matrix_multiplication $(BUILDDIR)/test-cnn_convolution
build-tests: prepare-tests $(TESTS_OBJ) $(BUILDDIR)/test-cnn_matrix_multiplication $(BUILDDIR)/test-cnn_convolution $(BUILDDIR)/test-cuda_memory_management
prepare-tests:
@rm -f $(BUILDDIR)/test-*
build/test-cnn_%: $(TEST_SRCDIR)/cnn_%.c $(CNN_OBJ) $(BUILDDIR)/colors.o $(BUILDDIR)/mnist.o $(BUILDDIR)/utils.o $(BUILDDIR)/memory_management.o
$(BUILDDIR)/test-cnn_%: $(TEST_SRCDIR)/cnn_%.c $(CNN_OBJ) $(BUILDDIR)/colors.o $(BUILDDIR)/mnist.o $(BUILDDIR)/utils.o $(BUILDDIR)/memory_management.o
$(CC) $^ -o $@ $(CFLAGS) $(LD_CFLAGS)
# mnist.o est déjà inclus en tant que mnist_mnist.o
build/test-mnist_%: $(TEST_SRCDIR)/mnist_%.c $(MNIST_OBJ) $(BUILDDIR)/colors.o
$(BUILDDIR)/test-mnist_%: $(TEST_SRCDIR)/mnist_%.c $(MNIST_OBJ) $(BUILDDIR)/colors.o
$(CC) $^ -o $@ $(CFLAGS) $(LD_CFLAGS)
build/test-memory_management: $(TEST_SRCDIR)/memory_management.c $(BUILDDIR)/colors.o $(BUILDDIR)/mnist.o $(BUILDDIR)/utils.o $(BUILDDIR)/test_memory_management.o
$(BUILDDIR)/test-memory_management: $(TEST_SRCDIR)/memory_management.c $(BUILDDIR)/colors.o $(BUILDDIR)/utils.o $(BUILDDIR)/test_memory_management.o
$(CC) $^ -o $@ $(CFLAGS) $(LD_CFLAGS)
$(BUILDDIR)/test_memory_management.o: $(SRCDIR)/memory_management.c $(SRCDIR)/include/memory_management.h
$(CC) -c $< -o $@ $(CFLAGS) -DTEST_MEMORY_MANAGEMENT
ifdef NVCC_INSTALLED
$(BUILDDIR)/test-cuda_memory_management: $(TEST_SRCDIR)/memory_management.cu $(BUILDDIR)/colors.cuda.o $(BUILDDIR)/cuda_utils.o $(BUILDDIR)/cuda_memory_management.o
$(NVCC) $(LD_NVCCFLAGS) $(NVCCFLAGS) $^ -o $@
else
$(BUILDDIR)/test-cuda_memory_management:
@echo "$(NVCC) not found, skipping"
endif
ifdef NVCC_INSTALLED
$(BUILDDIR)/test-cnn_%: $(TEST_SRCDIR)/cnn_%.cu \
$(BUILDDIR)/cnn_cuda_%.o \

6
src/include/memory_management.h
View File

@ -19,12 +19,18 @@ typedef struct Memory {
} Memory;
// Renvoie le nombre d'allocations totales dans la mémoire
#ifdef __CUDACC__
extern "C"
#endif
int get_memory_distinct_allocations();
// Fonction récursive correspondante
int get_distinct_allocations(Memory* mem);
// Renvoie le nombre d'éléments dans la liste chaînée représentant la mémoire
#ifdef __CUDACC__
extern "C"
#endif
int get_memory_blocks_number();
// Renvoie la taille d'une liste chaînée

92
test/memory_management.cu Normal file
View File

@ -0,0 +1,92 @@
#include <stdlib.h>
#include <stdio.h>
#include <assert.h>
#include "../src/include/memory_management.h"
#include "../src/include/colors.h"
#include "../src/include/utils.h"
#define N 350
__global__ void check_access(int* array, int range) {
for (int i=0; i < range; i++) {
array[i]++;
}
}
int main() {
printf("Pollution de la mémoire\n");
int mem_used;
int blocks_used;
// We pollute a little bit the memory before the tests
int* pointeurs[N];
for (int i=1; i < N; i++) {
pointeurs[i] = (int*)nalloc(i*sizeof(int));
for (int j=0; j < i; j++) {
pointeurs[i][j] = i;
}
}
// We test in a first place that one simple allocation works as expected
mem_used = get_memory_distinct_allocations();
blocks_used = get_memory_blocks_number();
void* ptr = nalloc(15);
if (! (get_memory_distinct_allocations() <= mem_used+1)) {
printf("Plus d'un élément de mémoire alloué en une seule allocation\n");
exit(1);
}
gree(ptr);
if (! (get_memory_blocks_number() == blocks_used)) {
printf("La mémoire n'a pas été libérée correctement\n");
exit(1);
}
printf(GREEN "OK\n" RESET);
printf("Vérification de l'accès CUDA\n");
/* On lance des kernels detaille 1 ce qui est itératif synchrone
* Donc un peu contraire à CUDA mais l'objectif est de débugger faiclement */
dim3 gridSize(1, 1, 1);
dim3 blockSize(1, 1, 1);
for (int i=1; i < N; i++) {
check_access<<<gridSize, blockSize>>>(pointeurs[i], i);
gpuErrchk( cudaPeekAtLastError() );
gpuErrchk( cudaDeviceSynchronize() );
}
printf(GREEN "OK\n" RESET);
printf("Allocation de deux demi-blocs\n");
// We test that we do not use too much blocks
blocks_used = get_memory_blocks_number();
void* ptr1 = nalloc(-1+MEMORY_BLOCK/2);
void* ptr2 = nalloc(-1+MEMORY_BLOCK/2);
if (! (get_memory_blocks_number() <= blocks_used +1)) {
printf("Trop de blocs ont été alloués par rapport à la mémoire nécessaire\n");
exit(1);
}
printf(GREEN "OK\n" RESET);
printf("Libération de la mémoire\n");
for (int i=1; i < N; i++) {
for (int j=0; j < i; j++) {
// We test that the memory does not overlap itself
assert(pointeurs[i][j] == i+1);
}
gree(pointeurs[i]);
}
gree(ptr1);
gree(ptr2);
if (! (get_memory_distinct_allocations() == 0 && get_memory_blocks_number() == 0)) {
printf("La mémoire n'a pas été libérée correctement\n");
exit(1);
}
printf(GREEN "OK\n" RESET);
return 0;
}