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

Commit 3104f90

Browse files
oluptonolupton
authored
Code generation changes for "inline" scopmath solvers. (#81)
* Use named structs instead of functions, pass instances to solvers. * Goes with BlueBrain/CoreNeuron#809.
1 parent d8c6e45 commit 3104f90

File tree

17 files changed

+114
-133
lines changed

17 files changed

+114
-133
lines changed

‎src/mod2c_core/deriv.c‎

Lines changed: 6 additions & 32 deletions
Original file line numberDiff line numberDiff line change
@@ -156,7 +156,7 @@ fun->name, listnum, maxerr_str);
156156
}else{
157157
Sprintf(buf,
158158
" if (!_thread[_spth%d]._pvoid) {\n"
159-
" _thread[_spth%d]._pvoid = nrn_cons_sparseobj(_kinetic_%s%s, %d, _ml, _threadargs_);\n"
159+
" _thread[_spth%d]._pvoid = nrn_cons_sparseobj(%s%s{}, %d, _ml, _threadargs_);\n"
160160
" #ifdef _OPENACC\n"
161161
" if (_nt->compute_gpu) {\n"
162162
" void* _d_so = (void*) acc_deviceptr(_thread[_spth%d]._pvoid);\n"
@@ -177,39 +177,13 @@ dindepname, fun->name, listnum, listnum);
177177
replacstr(qsol, buf);
178178
#if VECTORIZE
179179
if (method->subtype & DERF) { /* derivimplicit */
180-
Sprintf(buf,
181-
"\n"
182-
" #if !defined(_%s_%s%s)\n"
183-
" #define _%s_%s%s 0\n"
184-
" #endif\n"
185-
" %s%s_thread(%d, _slist%d, _dlist%d, _%s_%s%s, _threadargs_);\n",
186-
method->name, fun->name, suffix, method->name, fun->name,
187-
suffix, ssprefix, method->name,
188-
numeqn, listnum, listnum, method->name, fun->name, suffix);
180+
Sprintf(buf, "\n %s%s_thread(%d, _slist%d, _dlist%d, %s%s{}, _threadargs_);\n",
181+
ssprefix, method->name, numeqn, listnum, listnum, fun->name, suffix);
189182
vectorize_substitute(qsol, buf);
190-
191-
// euler_thread is defined externally and need a callback function
192-
// selection of callback is using switch-case implemented in _kinderiv.h
193-
if(strcmp(method->name, "euler") == 0) {
194-
Sprintf(buf,
195-
"\n"
196-
"/* _euler_ %s %s */\n"
197-
"#ifndef INSIDE_NMODL\n"
198-
"#define INSIDE_NMODL\n"
199-
"#endif\n"
200-
, fun->name, suffix);
201-
Linsertstr(procfunc, buf);
202-
}
203-
204183
}else{ /* kinetic */
205184
if (vectorize) {
206185
Sprintf(buf,
207-
"\n"
208-
" #if !defined(_kinetic_%s%s)\n"
209-
" #define _kinetic_%s%s 0\n"
210-
" #endif\n"
211-
" %s%s_thread((SparseObj*)_thread[_spth%d]._pvoid, %d, _slist%d, _dlist%d, &%s, %s, _kinetic_%s%s, _linmat%d, _threadargs_);\n",
212-
fun->name, suffix, fun->name, suffix,
186+
"\n %s%s_thread(static_cast<SparseObj*>(_thread[_spth%d]._pvoid), %d, _slist%d, _dlist%d, &%s, %s, %s%s{}, _linmat%d, _threadargs_);\n",
213187
ssprefix, method->name, listnum, numeqn, listnum, listnum, indepsym->name,
214188
dindepname, fun->name, suffix, listnum);
215189
vectorize_substitute(qsol, buf);
@@ -536,15 +510,15 @@ void massagederiv(q1, q2, q3, q4, sensused)
536510
/* all this junk is still in the intoken list */
537511
Sprintf(buf, "static inline int %s(_threadargsproto_);\n", SYM(q2)->name);
538512
if (deriv_implicit_really == 1) {
539-
Sprintf(buf, "extern int %s(_threadargsproto_);\n", SYM(q2)->name);
513+
Sprintf(buf, "struct %s%s {\n int operator()(_threadargsproto_) const;\n};\n", SYM(q2)->name, suffix);
540514
}
541515
Linsertstr(procfunc, buf);
542516
if (deriv_implicit_really == 1) {
543517
replacstr(q1, "\nint"); q = insertstr(q3, "() {_reset=0;\n");
544518
}else{
545519
replacstr(q1, "\nstatic int"); q = insertstr(q3, "() {_reset=0;\n");
546520
}
547-
vectorize_substitute(q, "(_threadargsproto_) {int _reset=0; int error = 0;\n");
521+
vectorize_substitute(q, "::operator()(_threadargsproto_) const {\n int _reset=0;\n int error = 0;\n");
548522

549523
if (derfun->subtype & DERF && derfun->u.i) {
550524
diag("DERIVATIVE merging not implemented", (char *)0);

‎src/mod2c_core/kinetic.c‎

Lines changed: 5 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -357,13 +357,16 @@ void massagekinetic(q1, q2, q3, q4, sensused) /*KINETIC NAME stmtlist '}'*/
357357
"#endif\n"
358358
, SYM(q2)->name, suffix);
359359
Linsertstr(procfunc, buf);
360-
replacstr(q1, "\nint");
360+
Sprintf(buf, "\nstruct %s%s {\n int operator()(SparseObj* _so, double* _rhs, _threadargsproto_) const;\n};\nint", SYM(q2)->name, suffix);
361+
replacstr(q1, buf);
362+
Sprintf(buf, "%s%s::operator()", SYM(q2)->name, suffix);
363+
replacstr(q2, buf);
361364
#endif
362365
qv = insertstr(q3, "()\n");
363366
#if VECTORIZE
364367
if (vectorize) {
365368
kin_vect1(q1, q2, q4);
366-
vectorize_substitute(qv, "(void* _so, double* _rhs, _threadargsproto_)\n");
369+
vectorize_substitute(qv, "(SparseObj* _so, double* _rhs, _threadargsproto_) const\n");
367370
}
368371
#endif
369372
qv = insertstr(q3, "{_reset=0;\n");

‎src/mod2c_core/nocpout.c‎

Lines changed: 3 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -343,11 +343,12 @@ fprintf(stderr, "Notice: ARTIFICIAL_CELL models that would require thread specif
343343
\n#include \"coreneuron/utils/ivocvect.hpp\"\
344344
\n#include \"coreneuron/utils/nrnoc_aux.hpp\"\
345345
\n#include \"coreneuron/gpu/nrn_acc_manager.hpp\"\
346-
\n#include \"coreneuron/mechanism/mech/cfile/scoplib.h\"\n\
347346
\n#include \"coreneuron/sim/scopmath/newton_struct.h\"\
347+
\n#include \"coreneuron/sim/scopmath/newton_thread.hpp\"\
348+
\n#include \"coreneuron/sim/scopmath/sparse_thread.hpp\"\
349+
\n#include \"coreneuron/sim/scopmath/ssimplic_thread.hpp\"\
348350
\n#include \"coreneuron/nrnoc/md2redef.h\"\
349351
\n#include \"coreneuron/mechanism/register_mech.hpp\"\
350-
\n#include \"_kinderiv.h\"\
351352
\n#if !NRNGPU\
352353
\n#if !defined(DISABLE_HOC_EXP)\
353354
\n#undef exp\

‎src/mod2c_core/simultan.c‎

Lines changed: 3 additions & 8 deletions
Original file line numberDiff line numberDiff line change
@@ -234,17 +234,12 @@ numlist, numlist-1, counts);
234234
counts, numlist, SYM(q2)->name, suffix, numlist);
235235
qret = insertstr(q3, buf);
236236
Sprintf(buf,
237-
"#pragma acc routine(nrn_newton_thread) seq\n"
238-
#if 0
239-
"_reset = nrn_newton_thread(_newtonspace%d, %d,_slist%d, _newton_%s%s, _dlist%d, _threadargs_);\n"
240-
#else
241-
"_reset = nrn_newton_thread((NewtonSpace*)_newtonspace%d, %d,_slist%d, _derivimplicit_%s%s, _dlist%d, _threadargs_);\n"
242-
#endif
237+
"_reset = nrn_newton_thread(static_cast<NewtonSpace*>(_newtonspace%d), %d, _slist%d, _newton_%s%s{}, _dlist%d, _threadargs_);\n"
243238
, numlist-1, counts, numlist, SYM(q2)->name, suffix, numlist);
244239
vectorize_substitute(qret, buf);
245240
Insertstr(q3, "/*if(_reset) {abort_run(_reset);}*/ }\n");
246241
Sprintf(buf,
247-
"int _newton_%s%s(_threadargsproto_);\n"
242+
"\nstruct _newton_%s%s {\n int operator()(_threadargsproto_) const;\n};\n"
248243
, SYM(q2)->name, suffix);
249244
Linsertstr(procfunc, buf);
250245

@@ -257,7 +252,7 @@ numlist, numlist-1, counts);
257252
, SYM(q2)->name, suffix);
258253
Linsertstr(procfunc, buf);
259254

260-
Sprintf(buf, "\n return _reset;\n}\n\nint _newton_%s%s (_threadargsproto_) { int _reset=0;\n", SYM(q2)->name, suffix);
255+
Sprintf(buf, "\n return _reset;\n}\n\nint _newton_%s%s::operator()(_threadargsproto_) const {\n int _reset=0;\n", SYM(q2)->name, suffix);
261256
Insertstr(q3, buf);
262257
q = insertstr(q3, "{ int _counte = -1;\n");
263258
sprintf(buf, "{ double* _savstate%d = (double*)_thread[_dith%d]._pval;\n\

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

Lines changed: 3 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -13,12 +13,12 @@
1313
#include "coreneuron/utils/ivocvect.hpp"
1414
#include "coreneuron/utils/nrnoc_aux.hpp"
1515
#include "coreneuron/gpu/nrn_acc_manager.hpp"
16-
#include "coreneuron/mechanism/mech/cfile/scoplib.h"
17-
1816
#include "coreneuron/sim/scopmath/newton_struct.h"
17+
#include "coreneuron/sim/scopmath/newton_thread.hpp"
18+
#include "coreneuron/sim/scopmath/sparse_thread.hpp"
19+
#include "coreneuron/sim/scopmath/ssimplic_thread.hpp"
1920
#include "coreneuron/nrnoc/md2redef.h"
2021
#include "coreneuron/mechanism/register_mech.hpp"
21-
#include "_kinderiv.h"
2222
#if !NRNGPU
2323
#if !defined(DISABLE_HOC_EXP)
2424
#undef exp

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

Lines changed: 3 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -13,12 +13,12 @@
1313
#include "coreneuron/utils/ivocvect.hpp"
1414
#include "coreneuron/utils/nrnoc_aux.hpp"
1515
#include "coreneuron/gpu/nrn_acc_manager.hpp"
16-
#include "coreneuron/mechanism/mech/cfile/scoplib.h"
17-
1816
#include "coreneuron/sim/scopmath/newton_struct.h"
17+
#include "coreneuron/sim/scopmath/newton_thread.hpp"
18+
#include "coreneuron/sim/scopmath/sparse_thread.hpp"
19+
#include "coreneuron/sim/scopmath/ssimplic_thread.hpp"
1920
#include "coreneuron/nrnoc/md2redef.h"
2021
#include "coreneuron/mechanism/register_mech.hpp"
21-
#include "_kinderiv.h"
2222
#if !NRNGPU
2323
#if !defined(DISABLE_HOC_EXP)
2424
#undef exp

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

Lines changed: 17 additions & 13 deletions
Original file line numberDiff line numberDiff line change
@@ -13,12 +13,12 @@
1313
#include "coreneuron/utils/ivocvect.hpp"
1414
#include "coreneuron/utils/nrnoc_aux.hpp"
1515
#include "coreneuron/gpu/nrn_acc_manager.hpp"
16-
#include "coreneuron/mechanism/mech/cfile/scoplib.h"
17-
1816
#include "coreneuron/sim/scopmath/newton_struct.h"
17+
#include "coreneuron/sim/scopmath/newton_thread.hpp"
18+
#include "coreneuron/sim/scopmath/sparse_thread.hpp"
19+
#include "coreneuron/sim/scopmath/ssimplic_thread.hpp"
1920
#include "coreneuron/nrnoc/md2redef.h"
2021
#include "coreneuron/mechanism/register_mech.hpp"
21-
#include "_kinderiv.h"
2222
#if !NRNGPU
2323
#if !defined(DISABLE_HOC_EXP)
2424
#undef exp
@@ -283,7 +283,10 @@ static int _ode_spec1(_threadargsproto_);
283283
#ifndef INSIDE_NMODL
284284
#define INSIDE_NMODL
285285
#endif
286-
int _newton_states_NapDA(_threadargsproto_);
286+
287+
struct _newton_states_NapDA {
288+
int operator()(_threadargsproto_) const;
289+
};
287290

288291
#define _slist2 _slist2_NapDA
289292
int* _slist2;
@@ -296,7 +299,9 @@ int* _slist1;
296299
#define _dlist1 _dlist1_NapDA
297300
int* _dlist1;
298301
#pragma acc declare create(_dlist1)
299-
extern int states(_threadargsproto_);
302+
struct states_NapDA {
303+
int operator()(_threadargsproto_) const;
304+
};
300305

301306
/*CVODE*/
302307
static int _ode_spec1 (_threadargsproto_) {int _reset = 0; {
@@ -312,18 +317,20 @@ int* _dlist1;
312317
}
313318
/*END CVODE*/
314319

315-
int states (_threadargsproto_) {int _reset=0; int error = 0;
320+
int states ::operator()(_threadargsproto_) const {
321+
int _reset=0;
322+
int error = 0;
316323
{ double* _savstate1 = (double*)_thread[_dith1]._pval;
317324
double* _dlist2 = (double*)(_thread[_dith1]._pval) + (2*_cntml_padded);
318325
{int _id; for(_id=0; _id < 2; _id++) { _savstate1[_id*_STRIDE] = _p[_slist1[_id]*_STRIDE];}}
319-
#pragma acc routine(nrn_newton_thread) seq
320-
_reset = nrn_newton_thread((NewtonSpace*)_newtonspace1, 2,_slist2, _derivimplicit_states_NapDA, _dlist2, _threadargs_);
326+
_reset = nrn_newton_thread(static_cast<NewtonSpace*>(_newtonspace1), 2, _slist2, _newton_states_NapDA{}, _dlist2, _threadargs_);
321327
/*if(_reset) {abort_run(_reset);}*/ }
322328

323329
return _reset;
324330
}
325331

326-
int _newton_states_NapDA (_threadargsproto_) { int _reset=0;
332+
int _newton_states_NapDA::operator()(_threadargsproto_) const {
333+
int _reset=0;
327334
{ double* _savstate1 = (double*)_thread[_dith1]._pval;
328335
double* _dlist2 = (double*)(_thread[_dith1]._pval) + (2*_cntml_padded);
329336
int _counte = -1;
@@ -620,10 +627,7 @@ for (;;) { /* help clang-format properly indent */
620627
v=_v;
621628
{
622629
{
623-
#if !defined(_derivimplicit_states_NapDA)
624-
#define _derivimplicit_states_NapDA 0
625-
#endif
626-
derivimplicit_thread(2, _slist1, _dlist1, _derivimplicit_states_NapDA, _threadargs_);
630+
derivimplicit_thread(2, _slist1, _dlist1, states_NapDA{}, _threadargs_);
627631
} }}
628632

629633
}

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

Lines changed: 3 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -13,12 +13,12 @@
1313
#include "coreneuron/utils/ivocvect.hpp"
1414
#include "coreneuron/utils/nrnoc_aux.hpp"
1515
#include "coreneuron/gpu/nrn_acc_manager.hpp"
16-
#include "coreneuron/mechanism/mech/cfile/scoplib.h"
17-
1816
#include "coreneuron/sim/scopmath/newton_struct.h"
17+
#include "coreneuron/sim/scopmath/newton_thread.hpp"
18+
#include "coreneuron/sim/scopmath/sparse_thread.hpp"
19+
#include "coreneuron/sim/scopmath/ssimplic_thread.hpp"
1920
#include "coreneuron/nrnoc/md2redef.h"
2021
#include "coreneuron/mechanism/register_mech.hpp"
21-
#include "_kinderiv.h"
2222
#if !NRNGPU
2323
#if !defined(DISABLE_HOC_EXP)
2424
#undef exp

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

Lines changed: 3 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -13,12 +13,12 @@
1313
#include "coreneuron/utils/ivocvect.hpp"
1414
#include "coreneuron/utils/nrnoc_aux.hpp"
1515
#include "coreneuron/gpu/nrn_acc_manager.hpp"
16-
#include "coreneuron/mechanism/mech/cfile/scoplib.h"
17-
1816
#include "coreneuron/sim/scopmath/newton_struct.h"
17+
#include "coreneuron/sim/scopmath/newton_thread.hpp"
18+
#include "coreneuron/sim/scopmath/sparse_thread.hpp"
19+
#include "coreneuron/sim/scopmath/ssimplic_thread.hpp"
1920
#include "coreneuron/nrnoc/md2redef.h"
2021
#include "coreneuron/mechanism/register_mech.hpp"
21-
#include "_kinderiv.h"
2222
#if !NRNGPU
2323
#if !defined(DISABLE_HOC_EXP)
2424
#undef exp

‎test/validation/mod2c_core/cpp/Nap_No.cpp‎

Lines changed: 3 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -13,12 +13,12 @@
1313
#include "coreneuron/utils/ivocvect.hpp"
1414
#include "coreneuron/utils/nrnoc_aux.hpp"
1515
#include "coreneuron/gpu/nrn_acc_manager.hpp"
16-
#include "coreneuron/mechanism/mech/cfile/scoplib.h"
17-
1816
#include "coreneuron/sim/scopmath/newton_struct.h"
17+
#include "coreneuron/sim/scopmath/newton_thread.hpp"
18+
#include "coreneuron/sim/scopmath/sparse_thread.hpp"
19+
#include "coreneuron/sim/scopmath/ssimplic_thread.hpp"
1920
#include "coreneuron/nrnoc/md2redef.h"
2021
#include "coreneuron/mechanism/register_mech.hpp"
21-
#include "_kinderiv.h"
2222
#if !NRNGPU
2323
#if !defined(DISABLE_HOC_EXP)
2424
#undef exp

0 commit comments

Comments
 (0)