CUDA 4.0 and gcc 4.4.5 compatibility fixes

mjuric · Jan 30, 2012 · ca08498 · ca08498
1 parent edb899a
commit ca08498
Show file tree

Hide file tree

Showing 4 changed files with 11 additions and 10 deletions.
diff --git a/src/common/cux.cpp b/src/common/cux.cpp
@@ -144,7 +144,8 @@ void cuxSmartPtr_impl_t::gc()
 	}
 	else
 	{
-		cuxErrCheck( cudaFree(slave) );
+		if(slave != NULL)
+			cuxErrCheck( cudaFree(slave) );
 	}
 	slave = NULL;
 
@@ -161,11 +162,11 @@ void cuxSmartPtr_impl_t::gc()
 	}
 }
 
-unsigned cuxGetFreeMem(unsigned *totalptr = NULL)
+size_t cuxGetFreeMem(unsigned *totalptr = NULL)
 {
 #if !CUDA_DEVEMU
 	// Memory info
-	unsigned free = 0, total = 0;
+	size_t free = 0, total = 0;
 	cuxErrCheck( (cudaError)cuMemGetInfo(&free, &total) );
 	if(totalptr) *totalptr = total;
 	return free;
@@ -759,7 +760,7 @@ bool cux_init()
 
 #if !CUDA_DEVEMU
 	// Memory info
-	unsigned free = 0, total = 0;
+	size_t free = 0, total = 0;
 	cuxErrCheck( (cudaError)cuMemGetInfo(&free, &total) );
 	MLOG(verb2) << "Device memory (free, total): " << free / (1<<20) << "M, " << total / (1<<20) << "M";
 #endif

diff --git a/src/common/cux.h b/src/common/cux.h
@@ -716,8 +716,8 @@ template<int dim>
 	struct cuxTexCoords
 	{
 		afloat2 tc[dim];
-		__host__ const float2 &operator[](int i) const { return tc[i]; }
-		__host__       float2 &operator[](int i)       { return tc[i]; }
+		__device__ __host__ const float2 &operator[](int i) const { return tc[i]; }
+		__device__ __host__       float2 &operator[](int i)       { return tc[i]; }
 	};
 
 /**

diff --git a/src/modules/module_lib.h b/src/modules/module_lib.h
@@ -107,8 +107,8 @@
 		__device__ __host__ double lon() const { return atan2(sl, cl); }
 		__device__ __host__ double lat() const { return atan2(sb, cb); }
 
-		direction() {}
-		direction(double l_, double b_)				// l,b are in radians
+		__device__ __host__ direction() {}
+		__device__ __host__ direction(double l_, double b_)				// l,b are in radians
 		: cl(cos(l_)), cb(cos(b_)), sl(sin(l_)), sb(sin(b_))
 		{ }
 	};

diff --git a/src/skygen/skygen.h b/src/skygen/skygen.h
@@ -96,8 +96,8 @@ struct ALIGN(16) pencilBeam : public direction
 
 	int extIdx;		// index into per-beam extinction texture
 
-	pencilBeam() {}
-	pencilBeam(Radians l_, Radians b_, float X_, float Y_, int projIdx_, float dx_, float coveredFraction_, int extIdx_)
+	__device__ __host__ pencilBeam() {}
+	__device__ __host__ pencilBeam(Radians l_, Radians b_, float X_, float Y_, int projIdx_, float dx_, float coveredFraction_, int extIdx_)
 	: direction(l_, b_), X(X_), Y(Y_), projIdx(projIdx_), dx(dx_), dA(dx_*dx_), coveredFraction(coveredFraction_), extIdx(extIdx_)
 	{ }
 };