Skip to content
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

Linker-imposed model complexity limit on Windows #408

Closed
neworderofjamie opened this issue Mar 26, 2021 · 0 comments · Fixed by #602
Closed

Linker-imposed model complexity limit on Windows #408

neworderofjamie opened this issue Mar 26, 2021 · 0 comments · Fixed by #602

Comments

@neworderofjamie
Copy link
Contributor

#407 allows large models to be compiled on Windows if you have sufficient RAM but, trying to link the multi-area model code now fails with "LINK : fatal error LNK1189: library limit of 65535 objects exceeded". As the multi-area model has about 70000 synapse populations and, even with all the runner-size-reduction optimizations in place, there is still more than one symbol exported for each one so this is hardly surprising. Seemingly this isn't even a Visual Studio limitation but rather a limitation of the Windows PE executable format which, somewhere, uses 16-bit indices for symbols (https://developercommunity.visualstudio.com/t/fix-msvc-65535-symbol-limit-in-lib-files-lnk1189/220174).

Obviously lol Windows but, while #286 has meant that while the size of kernels generated for complex models are kept manageable as long as the model's relatively homogeneous, the runner can still get really big. This means that, even on platforms where this particular limit doesn't apply, the compilation time and memory requirements for huge runners currently limits the maximum model size (in terms of number of populations). Generating runner code that looks more like the following would be one option but quite how it would handle heterogeneous populations is another question:

enum NeuronPopulations
{
    NeuronPopulationPre,
    NeuronPopulationPost,
    NeuronPopulationMax,
}

scalar *V[NeuronPopulationMax];
scalar *d_V[NeuronPopulationMax];
const size_t popSizes[NeuronPopulationMax] = {100, 100};

void pushVToDevice(NeuronPopulations pop) {
    CHECK_CUDA_ERRORS(cudaMemcpy(d_V[pop], V[pop], popSizes[pop] * sizeof(scalar), cudaMemcpyHostToDevice));
}
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Projects
None yet
Development

Successfully merging a pull request may close this issue.

1 participant