Now making compiler flags customizable through brian preferences and …

…implemented the interface to GeNN "cpu_only" mode. This canbe useful on non-CUDA computers.

brian2genn/correctness_testing.py

@@ -12,7 +12,7 @@
 class GeNNConfiguration(Configuration):
     name = 'GeNN'
     def before_run(self):
-        brian2.prefs.reset_to_defaults()
+#        brian2.prefs.reset_to_defaults()
         brian2.prefs.codegen.loop_invariant_optimisations = False
         brian2.set_device('genn')
 

brian2genn/device.py

@@ -877,6 +877,7 @@ def build(self, directory='output', compile=True, run=True, use_GPU=True,
                                                         model_name= self.model_name,
                                                         ROOTDIR= os.path.abspath(directory),
                                                         source_files= self.source_files,
+                                                        prefs=prefs
                                                         ) 
             open(os.path.join(directory, 'WINmakefile'), 'w').write(Makefile_tmp)
         else:
@@ -885,6 +886,7 @@ def build(self, directory='output', compile=True, run=True, use_GPU=True,
                                                         model_name= self.model_name,
                                                         ROOTDIR= os.path.abspath(directory),
                                                         source_files= self.source_files,
+                                                        prefs= prefs
                                                         ) 
             open(os.path.join(directory, 'GNUmakefile'), 'w').write(Makefile_tmp)
 
@@ -893,6 +895,10 @@ def build(self, directory='output', compile=True, run=True, use_GPU=True,
         if compile:
             with std_silent(debug):
                 if os.sys.platform == 'win32':
+                    # copy .cu file of cpu_only is desired
+                    if prefs.devices.genn.cpu_only:
+                        cmd= "copy runner.cu runnner_cpu_only.cc"  
+                        call(cmd, cwd=directory) 
                     bitversion= ''
                     if os.getenv('PROCESSOR_ARCHITECTURE') == "AMD64":
                         bitversion= 'x86_amd64'
@@ -904,12 +910,23 @@ def build(self, directory='output', compile=True, run=True, use_GPU=True,
                     # Users are required to set their path to "Visual Studio/VC", e.g.
                     # setx VS_PATH "C:\Program Files (x86)\Microsoft Visual Studio 10.0"
                     cmd= "\""+os.getenv('VS_PATH')+"\\VC\\vcvarsall.bat\" " + bitversion
-                    cmd= cmd+" && buildmodel.bat "+self.model_name + " && nmake /f WINmakefile clean && nmake /f WINmakefile"
+                    cmd= cmd+" && buildmodel.bat "+self.model_name + " 0 ";
+                    if prefs.devices.genn.cpu_only:
+                        cmd+= "cpu_only "
+                    cmd+= "&& nmake /f WINmakefile clean && nmake /f WINmakefile"
+                    if prefs.devices.genn.cpu_only:
+                        cmd+= " CPU_ONLY=1"
                     call(cmd, cwd=directory)
                 else:
-                    call(["buildmodel.sh", self.model_name], cwd=directory)
-                    call(["make", "clean"], cwd=directory)
-                    call(["make"], cwd=directory)
+                    if prefs.devices.genn.cpu_only:
+                        shutil.copy(directory+"/runner.cu", directory+"/runner_cpu_only.cc")  
+                        call(["buildmodel.sh", self.model_name, "0", "cpu_only"], cwd=directory)
+                        call(["make", "clean"], cwd=directory)
+                        call(["make", "cpu_only"], cwd=directory)
+                    else:
+                        call(["buildmodel.sh", self.model_name], cwd=directory)
+                        call(["make", "clean"], cwd=directory)
+                        call(["make"], cwd=directory)
 
         if run:
             if not with_output:
@@ -919,6 +936,9 @@ def build(self, directory='output', compile=True, run=True, use_GPU=True,
                 stdout = None
                 stderr = None
             start_time = time.time()
+            if prefs.devices.genn.cpu_only and use_GPU:
+                logger.warn("Cannot use a GPU in cpu_only mode. Using CPU.")
+                use_GPU= False
             gpu_arg = "1" if use_GPU else "0"
             if gpu_arg == "1":
                 where= 'on GPU'

brian2genn/preferences.py

@@ -37,6 +37,21 @@ def valid_connectivity_decision(expr):
         docs= '''
         This preference allows users to set their own logical expression for the decision whether to use dense or sparse matrix methods based on Npre, Npost and Nsyn. Users should provide a valid Python expression that involves the variable sNpre (pre-synaptic neurons number), Npost (post-synaptic neuron number) and Nsyn (number of existing synapses) and return <True> for using dense matrices and <False> for using sparse. The default behaviour, e.g., corresponds to the logical expression 'Nsyn > 0.25*Npre*Npost'. ''',
         default= 'Nsyn > 0.25*Npre*Npost'
+    ),
+    unix_compiler_flags= BrianPreference(
+        docs= '''
+        This preference allows users to set their own compiler flags for teh eventual compilation of GeNN generated code. The flags will be applied both for the nvcc compilation and C++ compiler compilation (e.g. gcc, clang). This preference is for unix-based operating systems. ''',
+        default= '-O3 -ffast-math'
+    ),
+    windows_compiler_flags= BrianPreference(
+        docs= '''
+                This preference allows users to set their own compiler flags for teh eventual compilation of GeNN generated code. The flags will be applied both for the nvcc compilation and C++ compiler compilation (e.g. MCVC cl). This preference is for the windows operating system. ''',
+        default= '/O2'
+    ),
+    cpu_only= BrianPreference(
+        docs= '''
+        Set this preference to True if you want to compile in GeNN's "CPU only" mode. By default this flag is False and GeNN compiles both a GPU and a CPU version. Which one is run in that case is decided by the useGPU flag in the device.buld command invocation. You might want cpu_only mode if you want to use GeNN on a system that does not have CUDA installed or does not have hardware that supports CUDA.''',
+        default= False
     )
 )
 

brian2genn/run_correctness_testing.py

@@ -14,6 +14,8 @@
 prefs.codegen.loop_invariant_optimisations = False
 prefs.devices.genn.connectivity = 'AUTO'
 prefs.core.network.default_schedule= ['start', 'synapses', 'groups', 'thresholds', 'resets', 'end']
+prefs.devices.genn.unix_compiler_flags="-O1"
+prefs.devices.genn.cpu_only=True
 
 #c = GeNNConfiguration()
 #c.before_run()
@@ -31,7 +33,7 @@
 #print run_feature_tests(configurations=[DefaultConfiguration, 
 #                                         GeNNConfiguration],
 #                         feature_tests=[NeuronGroupIntegrationLinear]).tables_and_exceptions
-run_feature_tests(configurations=[DefaultConfiguration,GeNNConfiguration,CPPStandaloneConfiguration],feature_tests=[ 
+print run_feature_tests(configurations=[DefaultConfiguration,GeNNConfiguration,CPPStandaloneConfiguration],feature_tests=[ 
 #run_feature_tests(configurations=[GeNNConfiguration], feature_tests=[ 
 NeuronGroupIntegrationLinear, NeuronGroupIntegrationEuler, NeuronGroupLIF, NeuronGroupLIFRefractory, SynapsesPre, SynapsesPost,
 # SynapsesSTDPNoAutapse, 

brian2genn/templates/GNUmakefile

@@ -1,11 +1,16 @@
 ROOTDIR         := {{ROOTDIR}}
 EXECUTABLE      := runner
+{% if prefs['devices.genn.cpu_only'] %}
+SOURCES := runner_cpu_only.cc {% for source in source_files %} {{source}} {% endfor %}
+
+{% else %}
 SOURCES         := runner.cu {% for source in source_files %} {{source}} {% endfor %}
 
+{% endif %}
+
+
 INCLUDE_FLAGS   := -I.
-LINK_FLAGS      := 
-CCFLAGS         := 
-NVCCFLAGS       := 
+OPTIMIZATIONFLAGS := {{prefs['devices.genn.unix_compiler_flags']}}
 
 include $(GENN_PATH)/userproject/include/makefile_common_gnu.mk
 

brian2genn/templates/WINmakefile

@@ -1,6 +1,12 @@
 EXECUTABLE	= runner.exe
+!IF "$(CPU_ONLY)" != "1"
 SOURCES		= runner.cu {% for source in source_files %} {{source}} {% endfor %}
+!ELSE
+SOURCES		= runner_cpu_only.cc {% for source in source_files %} {{source}} {% endfor %}
+!ENDIF
 
 EXTRA_INCLUDE	= /I %CD%
+OPTIMIZATIONFLAGS = {{prefs.devices.genn.windows_compiler_flags}}
+
 
 !include $(GENN_PATH)\userproject\include\makefile_common_win.mk

brian2genn/templates/engine.cc

@@ -35,9 +35,11 @@ class engine
   void free_device_mem(); 
   void run(double, unsigned int); 
   void output_state(FILE *, unsigned int); 
+#ifndef CPU_ONLY
   void getStateFromGPU(); 
   void getSpikesFromGPU(); 
   void getSpikeNumbersFromGPU(); 
+#endif
   void output_spikes(FILE *, unsigned int); 
   void sum_spikes(); 
 };
@@ -76,11 +78,13 @@ engine::engine()
 
 void engine::init(unsigned int which)
 {
+#ifndef CPU_ONLY
   if (which == CPU) {
   }
   if (which == GPU) {
     copyStateToDevice();
   }
+#endif
 }
 
 //--------------------------------------------------------------------------
@@ -123,6 +127,7 @@ void engine::run(double runtime, //!< Duration of time to run the model for
       _run_{{sm.name}}_codeobject();
       {% endif %}
       {% endfor %}
+#ifndef CPU_ONLY
       if (which == GPU) {
 	  stepTimeGPU();
 	  t= t-DT;
@@ -144,6 +149,7 @@ void engine::run(double runtime, //!< Duration of time to run the model for
 	  pull{{sm.monitored}}StateFromDevice();
 	  {% endfor %}
       }
+#endif
       if (which == CPU) {
 	  stepTimeCPU();
 	  t= t-DT;
@@ -190,9 +196,10 @@ void engine::output_state(FILE *f, //!< File handle for a file to write the mode
 			   unsigned int which //!< Flag determining whether using GPU or CPU only
 			   )
 {
+#ifndef CPU_ONLY
   if (which == GPU) 
     copyStateFromDevice();
-
+#endif
   fprintf(f, "%f ", t);
   {% for neuron_model in neuron_models %}
   for (int i= 0; i < model.neuronN[{{loop.index-1}}]; i++) {
@@ -213,6 +220,7 @@ void engine::output_state(FILE *f, //!< File handle for a file to write the mode
   fprintf(f,"\n");
 }
 
+#ifndef CPU_ONLY
 //--------------------------------------------------------------------------
 /*! \brief Method for copying all variables of the last time step from the GPU
  
@@ -250,6 +258,8 @@ void engine::getSpikeNumbersFromGPU()
   copySpikeNFromDevice();
 }
 
+#endif
+
 //--------------------------------------------------------------------------
 /*! \brief Method for writing the spikes occurred in the last time step to a file
  */

brian2genn/templates/runner.cc

@@ -105,7 +105,9 @@ int main(int argc, char *argv[])
   //-----------------------------------------------------------------
 
   eng.init(which);         // this includes copying g's for the GPU version
+#ifndef CPU_ONLY
   copyStateToDevice();
+#endif
 
   //------------------------------------------------------------------
   // output general parameters to output file and start the simulation
@@ -119,10 +121,12 @@ int main(int argc, char *argv[])
   fprintf(timef,"%f \n", timer.getElapsedTime());
 
   // get the final results from the GPU 
+#ifndef CPU_ONLY
   if (which == GPU) {
     eng.getStateFromGPU();
     eng.getSpikesFromGPU();
   }
+#endif
   // translate GeNN arrays back to synaptic arrays
   {% for synapses in synapse_models %}
   {% if synapses.connectivity == 'DENSE' %}
@@ -178,7 +182,13 @@ using namespace std;
 #include "hr_time.cpp"
 #include "utils.h" // for CHECK_CUDA_ERRORS
 
+#ifndef CPU_ONLY
 #include <cuda_runtime.h>
+#else
+#define __host__
+#define __device__
+#endif
+
 
 #ifndef RAND
 #define RAND(Y,X) Y = Y * 1103515245 +12345;X= (unsigned int)(Y >> 16) & 32767

docs_sphinx/introduction/preferences.rst

@@ -3,9 +3,27 @@ Brian2GeNN specific preferences
 
 Connectivity
 ------------------
+The preference ``prefs.devices.genn.conncetivity`` determines what
+connectivity scheme is used within GeNN to represent the connections
+between neurons. GeNN supports the use of full connectivity matrices
+('DENSE') or a representation where connections are represented with
+sparse matrix methods ('SPARSE'). When this preference is set to
+'AUTO', a separately defined logic expression determines wheter
+'DENSE' or 'SPARSE' are used for each individula Synapses population.
+This logic expression can be set with the
+``prefs.devices.genn.connectivity_decision`` preference.
+
 
 Connectivity_decision
 ----------------------
+The ``prefs.devices.genn.connectivity_decision`` contains a logical
+expression that determines whether to use 'DENSE' or 'SPARSE' matrix
+methods in GeNN for a given matrix. 'DENSE' is used when the condition
+evaluates to ``True`` and 'SPARSE' is used otherwise. It's standard
+value is ``Nsyn > 0.25*Npre*Npost``, i.e. 'DENSE' is used whenever the
+number of actually existing synaptic connections is larger than a
+quarter of all possible synaptic connections between the pre- and
+post-synaptic populations. 
 
 Compiler options
 -----------------