From ca084982bc0e07192c22ae62b16aa7794e984cd9 Mon Sep 17 00:00:00 2001
From: Mario Juric <mjuric@ias.edu>
Date: Mon, 30 Jan 2012 09:11:11 -0800
Subject: [PATCH] CUDA 4.0 and gcc 4.4.5 compatibility fixes

---
 src/common/cux.cpp       | 9 +++++----
 src/common/cux.h         | 4 ++--
 src/modules/module_lib.h | 4 ++--
 src/skygen/skygen.h      | 4 ++--
 4 files changed, 11 insertions(+), 10 deletions(-)

diff --git a/src/common/cux.cpp b/src/common/cux.cpp
index 9e60644..19a234b 100644
--- 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
index 9675c20..4432c91 100644
--- 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
index 8427b8c..21972e1 100644
--- 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
index 4e6f7f7..939f72e 100644
--- 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_)
 	{ }
 };