Skip to content

Commit 81d3031

Browse files
committed
fix GPU device deallocation with OpenMP
1 parent ef7ad96 commit 81d3031

File tree

4 files changed

+5
-13
lines changed

4 files changed

+5
-13
lines changed

simwave/kernel/backend/c_code/forward/constant_density/2d/wave.c

Lines changed: 2 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -53,11 +53,11 @@ double forward(f_type *u, f_type *velocity, f_type *damp,
5353

5454
#ifdef GPU_OPENMP
5555

56-
// select the device
56+
// select the device
5757
#ifdef DEVICEID
5858
omp_set_default_device(DEVICEID);
5959
#endif
60-
60+
6161
size_t shot_record_size = wavelet_size * num_receivers;
6262
size_t u_size = num_snapshots * domain_size;
6363

@@ -433,8 +433,6 @@ double forward(f_type *u, f_type *velocity, f_type *damp,
433433
#pragma omp target exit data map(delete: rec_points_values[:rec_points_values_size])
434434
#pragma omp target exit data map(delete: rec_points_values_offset[:num_receivers])
435435
#pragma omp target exit data map(delete: wavelet[:wavelet_size * wavelet_count])
436-
#pragma omp target exit data map(delete: receivers[:shot_record_size])
437-
#pragma omp target exit data map(delete: u[:u_size])
438436
#endif
439437

440438
// get the end time

simwave/kernel/backend/c_code/forward/constant_density/3d/wave.c

Lines changed: 2 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -54,11 +54,11 @@ double forward(f_type *u, f_type *velocity, f_type *damp,
5454

5555
#ifdef GPU_OPENMP
5656

57-
// select the device
57+
// select the device
5858
#ifdef DEVICEID
5959
omp_set_default_device(DEVICEID);
6060
#endif
61-
61+
6262
size_t shot_record_size = wavelet_size * num_receivers;
6363
size_t u_size = num_snapshots * domain_size;
6464

@@ -519,8 +519,6 @@ double forward(f_type *u, f_type *velocity, f_type *damp,
519519
#pragma omp target exit data map(delete: rec_points_values[:rec_points_values_size])
520520
#pragma omp target exit data map(delete: rec_points_values_offset[:num_receivers])
521521
#pragma omp target exit data map(delete: wavelet[:wavelet_size * wavelet_count])
522-
#pragma omp target exit data map(delete: receivers[:shot_record_size])
523-
#pragma omp target exit data map(delete: u[:u_size])
524522
#endif
525523

526524
// get the end time

simwave/kernel/backend/c_code/forward/variable_density/2d/wave.c

Lines changed: 1 addition & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -454,9 +454,7 @@ double forward(f_type *u, f_type *velocity, f_type *density, f_type *damp,
454454
#pragma omp target exit data map(delete: rec_points_interval[:rec_points_interval_size])
455455
#pragma omp target exit data map(delete: rec_points_values[:rec_points_values_size])
456456
#pragma omp target exit data map(delete: rec_points_values_offset[:num_receivers])
457-
#pragma omp target exit data map(delete: wavelet[:wavelet_size * wavelet_count])
458-
#pragma omp target exit data map(delete: receivers[:shot_record_size])
459-
#pragma omp target exit data map(delete: u[:u_size])
457+
#pragma omp target exit data map(delete: wavelet[:wavelet_size * wavelet_count])
460458
#endif
461459

462460
// get the end time

simwave/kernel/backend/c_code/forward/variable_density/3d/wave.c

Lines changed: 0 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -545,8 +545,6 @@ double forward(f_type *u, f_type *velocity, f_type *density, f_type *damp,
545545
#pragma omp target exit data map(delete: rec_points_values[:rec_points_values_size])
546546
#pragma omp target exit data map(delete: rec_points_values_offset[:num_receivers])
547547
#pragma omp target exit data map(delete: wavelet[:wavelet_size * wavelet_count])
548-
#pragma omp target exit data map(delete: receivers[:shot_record_size])
549-
#pragma omp target exit data map(delete: u[:u_size])
550548
#endif
551549

552550
// get the end time

0 commit comments

Comments
 (0)