Continuing efforts to fix compilation errors under Windows

platforms/cuda/src/CudaForceInfo.h

@@ -37,26 +37,28 @@ namespace OpenMM {
  * about the behavior and requirements of that force.
  */
 
-class OPENMM_EXPORT CudaForceInfo {
+class CudaForceInfo {
 public:
     CudaForceInfo() {
     }
+    virtual ~CudaForceInfo() {
+    }
     /**
      * Get whether or not two particles have identical force field parameters.
      */
-    virtual bool areParticlesIdentical(int particle1, int particle2);
+    virtual OPENMM_EXPORT bool areParticlesIdentical(int particle1, int particle2);
     /**
      * Get the number of particle groups defined by this force.
      */
-    virtual int getNumParticleGroups();
+    virtual OPENMM_EXPORT int getNumParticleGroups();
     /**
      * Get the list of particles in a particular group.
      */
-    virtual void getParticlesInGroup(int index, std::vector<int>& particles);
+    virtual OPENMM_EXPORT void getParticlesInGroup(int index, std::vector<int>& particles);
     /**
      * Get whether two particle groups are identical.
      */
-    virtual bool areGroupsIdentical(int group1, int group2);
+    virtual OPENMM_EXPORT bool areGroupsIdentical(int group1, int group2);
 };
 
 } // namespace OpenMM

platforms/cuda/src/kernels/cudaCompact.cu

@@ -183,7 +183,7 @@ __global__ void moveValidElementsScattered(const T* dgData,T* dgCompact,const un
     }
 }
 
-void planCompaction(compactionPlan& d,bool stageOutput) {
+void OPENMMCUDA_EXPORT planCompaction(compactionPlan& d,bool stageOutput) {
     int device;
     cudaGetDevice(&device);
     cudaDeviceProp deviceProp;
@@ -195,11 +195,11 @@ void planCompaction(compactionPlan& d,bool stageOutput) {
     d.valid = true;
 }
 
-void destroyCompactionPlan(compactionPlan& d) {
+void OPENMMCUDA_EXPORT destroyCompactionPlan(compactionPlan& d) {
     if (d.valid) cudaFree(d.dgBlockCounts);
 }
 
-int compactStream(const compactionPlan& d,T* dOut,const T* dIn,const unsigned int* dValid,size_t len,size_t* dNumValid) {
+int OPENMMCUDA_EXPORT compactStream(const compactionPlan& d,T* dOut,const T* dIn,const unsigned int* dValid,size_t len,size_t* dNumValid) {
     if (!d.valid) {
         return -1;
     }

platforms/cuda/src/kernels/cudaCompact.h

@@ -25,6 +25,8 @@
   License:      Public Domain
 */
 
+#include "windowsExportCuda.h"
+
 struct compactionPlan {
     bool valid;
     unsigned int* dgBlockCounts;
@@ -33,12 +35,12 @@ struct compactionPlan {
 };
 
 extern "C"
-void planCompaction(compactionPlan& d,bool stageOutput=true);
+void OPENMMCUDA_EXPORT planCompaction(compactionPlan& d,bool stageOutput=true);
 
 extern "C"
-void destroyCompactionPlan(compactionPlan& d);
+void OPENMMCUDA_EXPORT destroyCompactionPlan(compactionPlan& d);
 
 extern "C"
-int compactStream(const compactionPlan& d,unsigned int* dOut,const unsigned int* dIn,const unsigned int* dValid,size_t len,size_t* dNumValid);
+int OPENMMCUDA_EXPORT compactStream(const compactionPlan& d,unsigned int* dOut,const unsigned int* dIn,const unsigned int* dValid,size_t len,size_t* dNumValid);
 
 #endif // __OPENMM_CUDACOMPACT_H__

platforms/cuda/src/kernels/kCalculateObcGbsaBornSum.cu

@@ -133,7 +133,7 @@ void kReduceObcGbsaBornSum_kernel()
     }   
 }
 
-void kReduceObcGbsaBornSum(gpuContext gpu)
+void OPENMMCUDA_EXPORT kReduceObcGbsaBornSum(gpuContext gpu)
 {
 //    printf("kReduceObcGbsaBornSum\n");
     kReduceObcGbsaBornSum_kernel<<<gpu->sim.blocks, 384>>>();
@@ -141,7 +141,7 @@ void kReduceObcGbsaBornSum(gpuContext gpu)
     LAUNCHERROR("kReduceObcGbsaBornSum");
 }
 
-void kCalculateObcGbsaBornSum(gpuContext gpu)
+void OPENMMCUDA_EXPORT kCalculateObcGbsaBornSum(gpuContext gpu)
 {
   //  printf("kCalculateObcgbsaBornSum\n");
     switch (gpu->sim.nonbondedMethod)

platforms/cuda/src/kernels/kCalculateObcGbsaForces2.cu

@@ -99,7 +99,7 @@ void GetCalculateObcGbsaForces2Sim(gpuContext gpu)
 #define METHOD_NAME(a, b) a##PeriodicByWarp##b
 #include "kCalculateObcGbsaForces2.h"
 
-void kCalculateObcGbsaForces2(gpuContext gpu)
+void OPENMMCUDA_EXPORT kCalculateObcGbsaForces2(gpuContext gpu)
 {
     //printf("kCalculateObcGbsaForces2\n");
     switch (gpu->sim.nonbondedMethod)

platforms/cuda/src/kernels/kForces.cu

@@ -39,7 +39,7 @@ using namespace std;
 
 static __constant__ cudaGmxSimulation cSim;
 
-void SetForcesSim(gpuContext gpu)
+void OPENMMCUDA_EXPORT SetForcesSim(gpuContext gpu)
 {
     cudaError_t status;
     status = cudaMemcpyToSymbol(cSim, &gpu->sim, sizeof(cudaGmxSimulation));     
@@ -65,7 +65,7 @@ void kClearForces_kernel()
     }
 }
 
-void kClearForces(gpuContext gpu)
+void OPENMMCUDA_EXPORT kClearForces(gpuContext gpu)
 {
 //    printf("kClearForces\n");
     kClearForces_kernel<<<gpu->sim.blocks, 384>>>();
@@ -279,7 +279,7 @@ void kReduceForces_kernel()
     }   
 }
 
-void kReduceForces(gpuContext gpu)
+void OPENMMCUDA_EXPORT kReduceForces(gpuContext gpu)
 {
  //   printf("kReduceForces\n");
     kReduceForces_kernel<<<gpu->sim.blocks, gpu->sim.bsf_reduce_threads_per_block>>>();