Skip to content
This repository was archived by the owner on Jun 8, 2023. It is now read-only.

Commit 6b59295

Browse files
pramodkpramodk
committed
update tests
1 parent 16d6fe4 commit 6b59295

File tree

13 files changed

+477
-357
lines changed

13 files changed

+477
-357
lines changed

‎test/validation/mod2c_core/cpp/NaSm.cpp‎

Lines changed: 28 additions & 22 deletions
Original file line numberDiff line numberDiff line change
@@ -136,7 +136,7 @@
136136
static ThreadDatum* _extcall_thread;
137137
/* external NEURON variables */
138138
extern double celsius;
139-
139+
#define nrn_ghk(v, ci, co, z) nrn_ghk(v, ci, co, z, celsius)
140140
#if 0 /*BBCORE*/
141141
/* declaration of user functions */
142142
static void _hoc_rates(void);
@@ -264,7 +264,7 @@ static void _update_global_variables(NrnThread *_nt, Memb_list *_ml) {
264264
if(_nt == nullptr || _ml == nullptr) {
265265
return;
266266
}
267-
_global_variables_t* _global_variables = reinterpret_cast<_global_variables_t*>(_ml->instance);
267+
_global_variables_t* _global_variables = reinterpret_cast<_global_variables_t*>(_ml->global_variables);
268268
_global_variables->_ml_mechtype = _mechtype;
269269
_global_variables->celsius = celsius;
270270
_global_variables->Etemp = Etemp;
@@ -278,26 +278,26 @@ static void _update_global_variables(NrnThread *_nt, Memb_list *_ml) {
278278
_global_variables->m0 = m0;
279279
#ifdef CORENEURON_ENABLE_GPU
280280
if (_nt->compute_gpu) {
281-
auto* _d_global_vars = cnrn_target_copyin(_global_variables);
282-
auto* _d_ml = reinterpret_cast<Memb_list*>(acc_deviceptr(_ml));
283-
cnrn_target_memcpy_to_device(&(_d_ml->instance), (void**)&(_d_global_vars));
281+
auto* _d_global_variables = cnrn_target_copyin(_global_variables);
282+
auto* _d_ml = reinterpret_cast<Memb_list*>(acc_deviceptr(_ml));
283+
cnrn_target_memcpy_to_device(&(_d_ml->global_variables), (void**)&(_d_global_variables));
284284
}
285285
#endif
286286
}
287287

288-
#define _slist1 reinterpret_cast<_global_variables_t*>(_ml->instance)->_slist1
289-
#define _dlist1 reinterpret_cast<_global_variables_t*>(_ml->instance)->_dlist1
290-
#define celsius reinterpret_cast<_global_variables_t*>(_ml->instance)->celsius
291-
#define Etemp reinterpret_cast<_global_variables_t*>(_ml->instance)->Etemp
292-
#define Vtm reinterpret_cast<_global_variables_t*>(_ml->instance)->Vtm
293-
#define Vsm reinterpret_cast<_global_variables_t*>(_ml->instance)->Vsm
294-
#define ena reinterpret_cast<_global_variables_t*>(_ml->instance)->ena
295-
#define ktm reinterpret_cast<_global_variables_t*>(_ml->instance)->ktm
296-
#define ksm reinterpret_cast<_global_variables_t*>(_ml->instance)->ksm
297-
#define tom reinterpret_cast<_global_variables_t*>(_ml->instance)->tom
298-
#define delta_t reinterpret_cast<_global_variables_t*>(_ml->instance)->delta_t
299-
#define m0 reinterpret_cast<_global_variables_t*>(_ml->instance)->m0
300-
#define _ml_mechtype reinterpret_cast<_global_variables_t*>(_ml->instance)->_ml_mechtype
288+
#define _slist1 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->_slist1
289+
#define _dlist1 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->_dlist1
290+
#define celsius reinterpret_cast<_global_variables_t*>(_ml->global_variables)->celsius
291+
#define Etemp reinterpret_cast<_global_variables_t*>(_ml->global_variables)->Etemp
292+
#define Vtm reinterpret_cast<_global_variables_t*>(_ml->global_variables)->Vtm
293+
#define Vsm reinterpret_cast<_global_variables_t*>(_ml->global_variables)->Vsm
294+
#define ena reinterpret_cast<_global_variables_t*>(_ml->global_variables)->ena
295+
#define ktm reinterpret_cast<_global_variables_t*>(_ml->global_variables)->ktm
296+
#define ksm reinterpret_cast<_global_variables_t*>(_ml->global_variables)->ksm
297+
#define tom reinterpret_cast<_global_variables_t*>(_ml->global_variables)->tom
298+
#define delta_t reinterpret_cast<_global_variables_t*>(_ml->global_variables)->delta_t
299+
#define m0 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->m0
300+
#define _ml_mechtype reinterpret_cast<_global_variables_t*>(_ml->global_variables)->_ml_mechtype
301301

302302
static const char *modelname = "A slow Sodium current";
303303

@@ -374,11 +374,17 @@ double _v, v; int* _ni; int _iml, _cntml_padded, _cntml_actual;
374374
_cntml_actual = _ml->_nodecount;
375375
_cntml_padded = _ml->_nodecount_padded;
376376
_thread = _ml->_thread;
377-
if(_ml->instance) {
378-
free(_ml->instance);
379-
_ml->instance = nullptr;
377+
if(_ml->global_variables) {
378+
#ifdef _OPENACC
379+
if(acc_is_present(_ml->global_variables, sizeof(_global_variables_t))) {
380+
acc_delete(_ml->global_variables, sizeof(_global_variables_t));
381+
}
382+
#endif
383+
free(_ml->global_variables);
384+
_ml->global_variables = nullptr;
380385
}
381-
_ml->instance = malloc(sizeof(_global_variables_t));
386+
_ml->global_variables = malloc(sizeof(_global_variables_t));
387+
_ml->global_variables_size = sizeof(_global_variables_t);
382388
_initlists(_ml);
383389
_update_global_variables(_nt, _ml);
384390
double * _nt_data = _nt->_data;

‎test/validation/mod2c_core/cpp/Nap.cpp‎

Lines changed: 22 additions & 16 deletions
Original file line numberDiff line numberDiff line change
@@ -142,7 +142,7 @@
142142
static ThreadDatum* _extcall_thread;
143143
/* external NEURON variables */
144144
extern double celsius;
145-
145+
#define nrn_ghk(v, ci, co, z) nrn_ghk(v, ci, co, z, celsius)
146146
#if 0 /*BBCORE*/
147147
/* declaration of user functions */
148148
static void _hoc_trates(void);
@@ -265,28 +265,28 @@ static void _update_global_variables(NrnThread *_nt, Memb_list *_ml) {
265265
if(_nt == nullptr || _ml == nullptr) {
266266
return;
267267
}
268-
_global_variables_t* _global_variables = reinterpret_cast<_global_variables_t*>(_ml->instance);
268+
_global_variables_t* _global_variables = reinterpret_cast<_global_variables_t*>(_ml->global_variables);
269269
_global_variables->_ml_mechtype = _mechtype;
270270
_global_variables->celsius = celsius;
271271
_global_variables->eNa = eNa;
272272
_global_variables->delta_t = delta_t;
273273
_global_variables->m0 = m0;
274274
#ifdef CORENEURON_ENABLE_GPU
275275
if (_nt->compute_gpu) {
276-
auto* _d_global_vars = cnrn_target_copyin(_global_variables);
277-
auto* _d_ml = reinterpret_cast<Memb_list*>(acc_deviceptr(_ml));
278-
cnrn_target_memcpy_to_device(&(_d_ml->instance), (void**)&(_d_global_vars));
276+
auto* _d_global_variables = cnrn_target_copyin(_global_variables);
277+
auto* _d_ml = reinterpret_cast<Memb_list*>(acc_deviceptr(_ml));
278+
cnrn_target_memcpy_to_device(&(_d_ml->global_variables), (void**)&(_d_global_variables));
279279
}
280280
#endif
281281
}
282282

283-
#define _slist1 reinterpret_cast<_global_variables_t*>(_ml->instance)->_slist1
284-
#define _dlist1 reinterpret_cast<_global_variables_t*>(_ml->instance)->_dlist1
285-
#define celsius reinterpret_cast<_global_variables_t*>(_ml->instance)->celsius
286-
#define eNa reinterpret_cast<_global_variables_t*>(_ml->instance)->eNa
287-
#define delta_t reinterpret_cast<_global_variables_t*>(_ml->instance)->delta_t
288-
#define m0 reinterpret_cast<_global_variables_t*>(_ml->instance)->m0
289-
#define _ml_mechtype reinterpret_cast<_global_variables_t*>(_ml->instance)->_ml_mechtype
283+
#define _slist1 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->_slist1
284+
#define _dlist1 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->_dlist1
285+
#define celsius reinterpret_cast<_global_variables_t*>(_ml->global_variables)->celsius
286+
#define eNa reinterpret_cast<_global_variables_t*>(_ml->global_variables)->eNa
287+
#define delta_t reinterpret_cast<_global_variables_t*>(_ml->global_variables)->delta_t
288+
#define m0 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->m0
289+
#define _ml_mechtype reinterpret_cast<_global_variables_t*>(_ml->global_variables)->_ml_mechtype
290290

291291
static const char *modelname = "nap";
292292

@@ -362,11 +362,17 @@ double _v, v; int* _ni; int _iml, _cntml_padded, _cntml_actual;
362362
_cntml_actual = _ml->_nodecount;
363363
_cntml_padded = _ml->_nodecount_padded;
364364
_thread = _ml->_thread;
365-
if(_ml->instance) {
366-
free(_ml->instance);
367-
_ml->instance = nullptr;
365+
if(_ml->global_variables) {
366+
#ifdef _OPENACC
367+
if(acc_is_present(_ml->global_variables, sizeof(_global_variables_t))) {
368+
acc_delete(_ml->global_variables, sizeof(_global_variables_t));
369+
}
370+
#endif
371+
free(_ml->global_variables);
372+
_ml->global_variables = nullptr;
368373
}
369-
_ml->instance = malloc(sizeof(_global_variables_t));
374+
_ml->global_variables = malloc(sizeof(_global_variables_t));
375+
_ml->global_variables_size = sizeof(_global_variables_t);
370376
_initlists(_ml);
371377
_update_global_variables(_nt, _ml);
372378
double * _nt_data = _nt->_data;

‎test/validation/mod2c_core/cpp/NapDA.cpp‎

Lines changed: 29 additions & 19 deletions
Original file line numberDiff line numberDiff line change
@@ -135,7 +135,8 @@
135135
static int hoc_nrnpointerindex = -1;
136136
static ThreadDatum* _extcall_thread;
137137
/* external NEURON variables */
138-
138+
extern double celsius;
139+
#define nrn_ghk(v, ci, co, z) nrn_ghk(v, ci, co, z, celsius)
139140
#if 0 /*BBCORE*/
140141
/* declaration of user functions */
141142
static void _hoc_hbet(void);
@@ -261,6 +262,7 @@ static void nrn_alloc(double* _p, Datum* _ppvar, int _type) {
261262
int _slist1[2];
262263
int _dlist1[2];
263264
int _slist2[2];
265+
double celsius;
264266
double delta_t;
265267
double h0;
266268
double m0;
@@ -271,27 +273,29 @@ static void _update_global_variables(NrnThread *_nt, Memb_list *_ml) {
271273
if(_nt == nullptr || _ml == nullptr) {
272274
return;
273275
}
274-
_global_variables_t* _global_variables = reinterpret_cast<_global_variables_t*>(_ml->instance);
276+
_global_variables_t* _global_variables = reinterpret_cast<_global_variables_t*>(_ml->global_variables);
275277
_global_variables->_ml_mechtype = _mechtype;
278+
_global_variables->celsius = celsius;
276279
_global_variables->delta_t = delta_t;
277280
_global_variables->h0 = h0;
278281
_global_variables->m0 = m0;
279282
#ifdef CORENEURON_ENABLE_GPU
280283
if (_nt->compute_gpu) {
281-
auto* _d_global_vars = cnrn_target_copyin(_global_variables);
282-
auto* _d_ml = reinterpret_cast<Memb_list*>(acc_deviceptr(_ml));
283-
cnrn_target_memcpy_to_device(&(_d_ml->instance), (void**)&(_d_global_vars));
284+
auto* _d_global_variables = cnrn_target_copyin(_global_variables);
285+
auto* _d_ml = reinterpret_cast<Memb_list*>(acc_deviceptr(_ml));
286+
cnrn_target_memcpy_to_device(&(_d_ml->global_variables), (void**)&(_d_global_variables));
284287
}
285288
#endif
286289
}
287290

288-
#define _slist1 reinterpret_cast<_global_variables_t*>(_ml->instance)->_slist1
289-
#define _dlist1 reinterpret_cast<_global_variables_t*>(_ml->instance)->_dlist1
290-
#define _slist2 reinterpret_cast<_global_variables_t*>(_ml->instance)->_slist2
291-
#define delta_t reinterpret_cast<_global_variables_t*>(_ml->instance)->delta_t
292-
#define h0 reinterpret_cast<_global_variables_t*>(_ml->instance)->h0
293-
#define m0 reinterpret_cast<_global_variables_t*>(_ml->instance)->m0
294-
#define _ml_mechtype reinterpret_cast<_global_variables_t*>(_ml->instance)->_ml_mechtype
291+
#define _slist1 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->_slist1
292+
#define _dlist1 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->_dlist1
293+
#define _slist2 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->_slist2
294+
#define celsius reinterpret_cast<_global_variables_t*>(_ml->global_variables)->celsius
295+
#define delta_t reinterpret_cast<_global_variables_t*>(_ml->global_variables)->delta_t
296+
#define h0 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->h0
297+
#define m0 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->m0
298+
#define _ml_mechtype reinterpret_cast<_global_variables_t*>(_ml->global_variables)->_ml_mechtype
295299

296300
static const char *modelname = "";
297301

@@ -482,6 +486,19 @@ double _v, v; int* _ni; int _iml, _cntml_padded, _cntml_actual;
482486
_cntml_actual = _ml->_nodecount;
483487
_cntml_padded = _ml->_nodecount_padded;
484488
_thread = _ml->_thread;
489+
if(_ml->global_variables) {
490+
#ifdef _OPENACC
491+
if(acc_is_present(_ml->global_variables, sizeof(_global_variables_t))) {
492+
acc_delete(_ml->global_variables, sizeof(_global_variables_t));
493+
}
494+
#endif
495+
free(_ml->global_variables);
496+
_ml->global_variables = nullptr;
497+
}
498+
_ml->global_variables = malloc(sizeof(_global_variables_t));
499+
_ml->global_variables_size = sizeof(_global_variables_t);
500+
_initlists(_ml);
501+
_update_global_variables(_nt, _ml);
485502
_deriv1_advance = 0;
486503
#ifdef _OPENACC
487504
#pragma acc update device (_deriv1_advance) if (_nt->compute_gpu)
@@ -499,13 +516,6 @@ _thread = _ml->_thread;
499516
}
500517
#endif
501518
}
502-
if(_ml->instance) {
503-
free(_ml->instance);
504-
_ml->instance = nullptr;
505-
}
506-
_ml->instance = malloc(sizeof(_global_variables_t));
507-
_initlists(_ml);
508-
_update_global_variables(_nt, _ml);
509519
double * _nt_data = _nt->_data;
510520
double * _vec_v = _nt->_actual_v;
511521
int stream_id = _nt->stream_id;

‎test/validation/mod2c_core/cpp/NapIn.cpp‎

Lines changed: 23 additions & 17 deletions
Original file line numberDiff line numberDiff line change
@@ -142,7 +142,7 @@
142142
static ThreadDatum* _extcall_thread;
143143
/* external NEURON variables */
144144
extern double celsius;
145-
145+
#define nrn_ghk(v, ci, co, z) nrn_ghk(v, ci, co, z, celsius)
146146
#if 0 /*BBCORE*/
147147
/* declaration of user functions */
148148
static void _hoc_trates(void);
@@ -264,7 +264,7 @@ static void _update_global_variables(NrnThread *_nt, Memb_list *_ml) {
264264
if(_nt == nullptr || _ml == nullptr) {
265265
return;
266266
}
267-
_global_variables_t* _global_variables = reinterpret_cast<_global_variables_t*>(_ml->instance);
267+
_global_variables_t* _global_variables = reinterpret_cast<_global_variables_t*>(_ml->global_variables);
268268
_global_variables->_ml_mechtype = _mechtype;
269269
_global_variables->celsius = celsius;
270270
_global_variables->eNa = eNa;
@@ -273,21 +273,21 @@ static void _update_global_variables(NrnThread *_nt, Memb_list *_ml) {
273273
_global_variables->m0 = m0;
274274
#ifdef CORENEURON_ENABLE_GPU
275275
if (_nt->compute_gpu) {
276-
auto* _d_global_vars = cnrn_target_copyin(_global_variables);
277-
auto* _d_ml = reinterpret_cast<Memb_list*>(acc_deviceptr(_ml));
278-
cnrn_target_memcpy_to_device(&(_d_ml->instance), (void**)&(_d_global_vars));
276+
auto* _d_global_variables = cnrn_target_copyin(_global_variables);
277+
auto* _d_ml = reinterpret_cast<Memb_list*>(acc_deviceptr(_ml));
278+
cnrn_target_memcpy_to_device(&(_d_ml->global_variables), (void**)&(_d_global_variables));
279279
}
280280
#endif
281281
}
282282

283-
#define _slist1 reinterpret_cast<_global_variables_t*>(_ml->instance)->_slist1
284-
#define _dlist1 reinterpret_cast<_global_variables_t*>(_ml->instance)->_dlist1
285-
#define celsius reinterpret_cast<_global_variables_t*>(_ml->instance)->celsius
286-
#define eNa reinterpret_cast<_global_variables_t*>(_ml->instance)->eNa
287-
#define delta_t reinterpret_cast<_global_variables_t*>(_ml->instance)->delta_t
288-
#define h0 reinterpret_cast<_global_variables_t*>(_ml->instance)->h0
289-
#define m0 reinterpret_cast<_global_variables_t*>(_ml->instance)->m0
290-
#define _ml_mechtype reinterpret_cast<_global_variables_t*>(_ml->instance)->_ml_mechtype
283+
#define _slist1 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->_slist1
284+
#define _dlist1 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->_dlist1
285+
#define celsius reinterpret_cast<_global_variables_t*>(_ml->global_variables)->celsius
286+
#define eNa reinterpret_cast<_global_variables_t*>(_ml->global_variables)->eNa
287+
#define delta_t reinterpret_cast<_global_variables_t*>(_ml->global_variables)->delta_t
288+
#define h0 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->h0
289+
#define m0 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->m0
290+
#define _ml_mechtype reinterpret_cast<_global_variables_t*>(_ml->global_variables)->_ml_mechtype
291291

292292
static const char *modelname = "nap";
293293

@@ -367,11 +367,17 @@ double _v, v; int* _ni; int _iml, _cntml_padded, _cntml_actual;
367367
_cntml_actual = _ml->_nodecount;
368368
_cntml_padded = _ml->_nodecount_padded;
369369
_thread = _ml->_thread;
370-
if(_ml->instance) {
371-
free(_ml->instance);
372-
_ml->instance = nullptr;
370+
if(_ml->global_variables) {
371+
#ifdef _OPENACC
372+
if(acc_is_present(_ml->global_variables, sizeof(_global_variables_t))) {
373+
acc_delete(_ml->global_variables, sizeof(_global_variables_t));
374+
}
375+
#endif
376+
free(_ml->global_variables);
377+
_ml->global_variables = nullptr;
373378
}
374-
_ml->instance = malloc(sizeof(_global_variables_t));
379+
_ml->global_variables = malloc(sizeof(_global_variables_t));
380+
_ml->global_variables_size = sizeof(_global_variables_t);
375381
_initlists(_ml);
376382
_update_global_variables(_nt, _ml);
377383
double * _nt_data = _nt->_data;

‎test/validation/mod2c_core/cpp/Nap_E.cpp‎

Lines changed: 25 additions & 15 deletions
Original file line numberDiff line numberDiff line change
@@ -149,7 +149,8 @@
149149
static int hoc_nrnpointerindex = -1;
150150
static ThreadDatum* _extcall_thread;
151151
/* external NEURON variables */
152-
152+
extern double celsius;
153+
#define nrn_ghk(v, ci, co, z) nrn_ghk(v, ci, co, z, celsius)
153154
#if 0 /*BBCORE*/
154155
/* declaration of user functions */
155156
static void _hoc_rates(void);
@@ -260,6 +261,7 @@ static void nrn_alloc(double* _p, Datum* _ppvar, int _type) {
260261
struct _global_variables_t {
261262
int _slist1[2];
262263
int _dlist1[2];
264+
double celsius;
263265
double delta_t;
264266
double h0;
265267
double m0;
@@ -270,26 +272,28 @@ static void _update_global_variables(NrnThread *_nt, Memb_list *_ml) {
270272
if(_nt == nullptr || _ml == nullptr) {
271273
return;
272274
}
273-
_global_variables_t* _global_variables = reinterpret_cast<_global_variables_t*>(_ml->instance);
275+
_global_variables_t* _global_variables = reinterpret_cast<_global_variables_t*>(_ml->global_variables);
274276
_global_variables->_ml_mechtype = _mechtype;
277+
_global_variables->celsius = celsius;
275278
_global_variables->delta_t = delta_t;
276279
_global_variables->h0 = h0;
277280
_global_variables->m0 = m0;
278281
#ifdef CORENEURON_ENABLE_GPU
279282
if (_nt->compute_gpu) {
280-
auto* _d_global_vars = cnrn_target_copyin(_global_variables);
281-
auto* _d_ml = reinterpret_cast<Memb_list*>(acc_deviceptr(_ml));
282-
cnrn_target_memcpy_to_device(&(_d_ml->instance), (void**)&(_d_global_vars));
283+
auto* _d_global_variables = cnrn_target_copyin(_global_variables);
284+
auto* _d_ml = reinterpret_cast<Memb_list*>(acc_deviceptr(_ml));
285+
cnrn_target_memcpy_to_device(&(_d_ml->global_variables), (void**)&(_d_global_variables));
283286
}
284287
#endif
285288
}
286289

287-
#define _slist1 reinterpret_cast<_global_variables_t*>(_ml->instance)->_slist1
288-
#define _dlist1 reinterpret_cast<_global_variables_t*>(_ml->instance)->_dlist1
289-
#define delta_t reinterpret_cast<_global_variables_t*>(_ml->instance)->delta_t
290-
#define h0 reinterpret_cast<_global_variables_t*>(_ml->instance)->h0
291-
#define m0 reinterpret_cast<_global_variables_t*>(_ml->instance)->m0
292-
#define _ml_mechtype reinterpret_cast<_global_variables_t*>(_ml->instance)->_ml_mechtype
290+
#define _slist1 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->_slist1
291+
#define _dlist1 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->_dlist1
292+
#define celsius reinterpret_cast<_global_variables_t*>(_ml->global_variables)->celsius
293+
#define delta_t reinterpret_cast<_global_variables_t*>(_ml->global_variables)->delta_t
294+
#define h0 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->h0
295+
#define m0 reinterpret_cast<_global_variables_t*>(_ml->global_variables)->m0
296+
#define _ml_mechtype reinterpret_cast<_global_variables_t*>(_ml->global_variables)->_ml_mechtype
293297

294298
static const char *modelname = "";
295299

@@ -415,11 +419,17 @@ double _v, v; int* _ni; int _iml, _cntml_padded, _cntml_actual;
415419
_cntml_actual = _ml->_nodecount;
416420
_cntml_padded = _ml->_nodecount_padded;
417421
_thread = _ml->_thread;
418-
if(_ml->instance) {
419-
free(_ml->instance);
420-
_ml->instance = nullptr;
422+
if(_ml->global_variables) {
423+
#ifdef _OPENACC
424+
if(acc_is_present(_ml->global_variables, sizeof(_global_variables_t))) {
425+
acc_delete(_ml->global_variables, sizeof(_global_variables_t));
426+
}
427+
#endif
428+
free(_ml->global_variables);
429+
_ml->global_variables = nullptr;
421430
}
422-
_ml->instance = malloc(sizeof(_global_variables_t));
431+
_ml->global_variables = malloc(sizeof(_global_variables_t));
432+
_ml->global_variables_size = sizeof(_global_variables_t);
423433
_initlists(_ml);
424434
_update_global_variables(_nt, _ml);
425435
double * _nt_data = _nt->_data;

0 commit comments

Comments
 (0)