Create 2.0.4 release.

2008-10-06 20:16:02 +00:00
58 changed files with 1667 additions and 2245 deletions
--- a/CMakeLists.txt
+++ b/CMakeLists.txt
@ -1,4 +1,4 @@
-CMAKE_MINIMUM_REQUIRED(VERSION 2.6.0)
+CMAKE_MINIMUM_REQUIRED(VERSION 2.4.0)
 PROJECT(NV)
 ENABLE_TESTING()

@ -16,13 +16,6 @@ MESSAGE(STATUS "Setting optimal options")
 MESSAGE(STATUS "  Processor: ${NV_SYSTEM_PROCESSOR}")
 MESSAGE(STATUS "  Compiler Flags: ${CMAKE_CXX_FLAGS}")

-IF(NVTT_SHARED)
-	SET(NVCORE_SHARED TRUE)
-	SET(NVMATH_SHARED TRUE)
-	SET(NVIMAGE_SHARED TRUE)
-ENDIF(NVTT_SHARED)
-
-
 ADD_SUBDIRECTORY(src)

 IF(WIN32)
--- a/34
+++ b/34
@ -1,37 +1,3 @@
-NVIDIA Texture Tools version 2.0.8
- * Fix float to fixed image conversion. Patch provided by Alex Pfaffe. Fixes issue 121.
- * ColorBlock::isSingleColor compares only RGB channels. Fixes issue 115.
- * Fix cmake build in msvc. Fixes issue 111.
- * Better estimate principal component. Fixes issue 120.
-
-NVIDIA Texture Tools version 2.0.7
- * Output correct exit codes. Fixes issue 92.
- * Fix thread-safety errors. Fixes issue 90.
- * Add SIMD power method. Fixes issue 94.
- * Interact better with applications that already use CUDA.
- * Faster CPU compression.
-
-NVIDIA Texture Tools version 2.0.6
- * Fix dll version checking.
- * Detect CUDA 2.1 and future CUDA versions correctly.
- * Print CUDA detection message in nvcompress.
- * Select the fastest CUDA device.
- * Compile squish with -fPIC. Fixes issue 74.
- * Fix warnings under gcc 4.3.2.
- * Fix nvzoom option typo by Frank Richter. Fixes issue 81.
- * Do not use CUDA to compress small mipmaps. Fixes issue 76.
- * Compute mipmaps of semi-transparent images correctly.
- * Shutdown CUDA properly. Fixes issue 83.
- * Fix pixel format converions. Fixes issue 87.
- * Update single color compression tables. Fixes issue 85.
-
-NVIDIA Texture Tools version 2.0.5
- * Fix error in single color compressor. Fixes issue 66.
- * Detect mismatch between CUDA runtime and driver, and disable CUDA in that case.
- * Fix cmake files when compiling NVTT as a shared library.
- * When linking nvtt dynamically on unix, link all libraries dynamically.
- * Select fastest CUDA device.
-
 NVIDIA Texture Tools version 2.0.4
 * Fix error in RGB format output; reported by jonsoh. See issue 49.
 * Added support RGB format dithering by jonsoh. Fixes issue 50 and 51.
--- a/2
+++ b/2
@ -1 +1 @@
-2.0.8
+2.0.4
--- a/cmake/FindCUDA.cmake
+++ b/cmake/FindCUDA.cmake
@ -57,7 +57,7 @@ MARK_AS_ADVANCED (CUDA_FOUND CUDA_COMPILER CUDA_RUNTIME_LIBRARY)


 #SET(CUDA_OPTIONS "-ncfe")
-SET(CUDA_OPTIONS "--host-compilation=C")
+SET(CUDA_OPTIONS "")

 IF (CUDA_EMULATION)
 	SET (CUDA_OPTIONS "${CUDA_OPTIONS} -deviceemu")
--- a/cmake/FindOpenEXR.cmake
+++ b/cmake/FindOpenEXR.cmake
@ -53,7 +53,11 @@ ENDIF(ZLIB_FOUND)
 IF (OPENEXR_INCLUDE_PATH AND OPENEXR_IMATH_LIBRARY AND OPENEXR_ILMIMF_LIBRARY AND OPENEXR_IEX_LIBRARY AND OPENEXR_HALF_LIBRARY)
 	SET(OPENEXR_FOUND TRUE)
 	SET(OPENEXR_INCLUDE_PATHS ${OPENEXR_INCLUDE_PATH} CACHE STRING "The include paths needed to use OpenEXR")
-	SET(OPENEXR_LIBRARIES ${OPENEXR_IMATH_LIBRARY} ${OPENEXR_ILMIMF_LIBRARY} ${OPENEXR_IEX_LIBRARY} ${OPENEXR_HALF_LIBRARY} ${OPENEXR_ILMTHREAD_LIBRARY} ${ZLIB_LIBRARY} CACHE STRING "The libraries needed to use OpenEXR")
+	SET(OPENEXR_LIBRARIES ${OPENEXR_IMATH_LIBRARY} ${OPENEXR_ILMIMF_LIBRARY} ${OPENEXR_IEX_LIBRARY} ${OPENEXR_HALF_LIBRARY} ${ZLIB_LIBRARY} CACHE STRING "The libraries needed to use OpenEXR")
+
+	IF(OPENEXR_ILMTHREAD_LIBRARY)
+		SET(OPENEXR_LIBRARIES ${OPENEXR_LIBRARIES} ${OPENEXR_ILMTHREAD_LIBRARY})
+	ENDIF(OPENEXR_ILMTHREAD_LIBRARY)
 ENDIF (OPENEXR_INCLUDE_PATH AND OPENEXR_IMATH_LIBRARY AND OPENEXR_ILMIMF_LIBRARY AND OPENEXR_IEX_LIBRARY AND OPENEXR_HALF_LIBRARY)

 IF(OPENEXR_FOUND)
--- a/2
+++ b/2
@ -53,7 +53,7 @@ echo "-- Configuring nvidia-texture-tools "`cat VERSION`

 mkdir -p ./build
 cd ./build
-$CMAKE .. -DNVTT_SHARED=1 -DCMAKE_BUILD_TYPE=$build -DCMAKE_INSTALL_PREFIX=$prefix -G "Unix Makefiles" || exit 1
+$CMAKE .. -DCMAKE_BUILD_TYPE=$build -DCMAKE_INSTALL_PREFIX=$prefix -G "Unix Makefiles" || exit 1
 cd ..

 echo ""
--- a/project/vc8/nvcore/nvcore.vcproj
+++ b/project/vc8/nvcore/nvcore.vcproj
@ -281,10 +281,6 @@
 				RelativePath="..\..\..\src\nvcore\Debug.cpp"
 				>
 			</File>
-			<File
-				RelativePath="..\..\..\src\nvcore\Library.cpp"
-				>
-			</File>
 			<File
 				RelativePath="..\..\..\src\nvcore\Memory.cpp"
 				>
@ -319,10 +315,6 @@
 				RelativePath="..\..\..\src\nvcore\DefsVcWin32.h"
 				>
 			</File>
-			<File
-				RelativePath="..\..\..\src\nvcore\Library.h"
-				>
-			</File>
 			<File
 				RelativePath="..\..\..\src\nvcore\Memory.h"
 				>
--- a/project/vc8/nvtt/nvtt.rc
+++ b/project/vc8/nvtt/nvtt.rc
@ -53,8 +53,8 @@ END
 //

 VS_VERSION_INFO VERSIONINFO
- FILEVERSION 2,0,6,0
- PRODUCTVERSION 2,0,6,0
+ FILEVERSION 2,0,4,0
+ PRODUCTVERSION 2,0,4,0
 FILEFLAGSMASK 0x17L
 #ifdef _DEBUG
 FILEFLAGS 0x1L
@ -71,12 +71,12 @@ BEGIN
        BEGIN
            VALUE "CompanyName", "NVIDIA Corporation"
            VALUE "FileDescription", "NVIDIA Texture Tools Dynamic Link Library"
-            VALUE "FileVersion", "2, 0, 8, 0"
+            VALUE "FileVersion", "2, 0, 4, 0"
            VALUE "InternalName", "nvtt"
-            VALUE "LegalCopyright", "Copyright (C) 2007-2010"
+            VALUE "LegalCopyright", "Copyright (C) 2007"
            VALUE "OriginalFilename", "nvtt.dll"
            VALUE "ProductName", "NVIDIA Texture Tools Dynamic Link Library"
-            VALUE "ProductVersion", "2, 0, 8, 0"
+            VALUE "ProductVersion", "2, 0, 4, 0"
        END
    END
    BLOCK "VarFileInfo"
--- a/project/vc8/nvtt/nvtt.vcproj
+++ b/project/vc8/nvtt/nvtt.vcproj
@ -179,8 +179,6 @@
 			/>
 			<Tool
 				Name="VCPostBuildEventTool"
-				Description="Copying header files..."
-				CommandLine="xcopy /y /f /i &quot;$(SolutionDir)\..\..\src\nvtt\nvtt*.h&quot; &quot;$(SolutionDir)\$(ConfigurationName).$(PlatformName)\include\nvtt\&quot;"
 			/>
 		</Configuration>
 		<Configuration
@ -346,8 +344,6 @@
 			/>
 			<Tool
 				Name="VCPostBuildEventTool"
-				Description="Copying header files..."
-				CommandLine="xcopy /y /f /i &quot;$(SolutionDir)\..\..\src\nvtt\nvtt*.h&quot; &quot;$(SolutionDir)\$(ConfigurationName).$(PlatformName)\include\nvtt\&quot;"
 			/>
 		</Configuration>
 		<Configuration
@ -507,8 +503,6 @@
 			/>
 			<Tool
 				Name="VCPostBuildEventTool"
-				Description="Copying header files..."
-				CommandLine="xcopy /y /f /i &quot;$(SolutionDir)\..\..\src\nvtt\nvtt*.h&quot; &quot;$(SolutionDir)\$(ConfigurationName).$(PlatformName)\include\nvtt\&quot;"
 			/>
 		</Configuration>
 		<Configuration
@ -670,8 +664,6 @@
 			/>
 			<Tool
 				Name="VCPostBuildEventTool"
-				Description="Copying header files..."
-				CommandLine="xcopy /y /f /i &quot;$(SolutionDir)\..\..\src\nvtt\nvtt*.h&quot; &quot;$(SolutionDir)\$(ConfigurationName).$(PlatformName)\include\nvtt\&quot;"
 			/>
 		</Configuration>
 	</Configurations>
--- a/src/CMakeLists.txt
+++ b/src/CMakeLists.txt
@ -105,8 +105,7 @@ ENDIF(OPENEXR_FOUND)
 FIND_PACKAGE(Qt4)

 # Threads
-FIND_PACKAGE(Threads REQUIRED)
-MESSAGE(STATUS "Use thread library: ${CMAKE_THREAD_LIBS_INIT}")
+FIND_PACKAGE(Threads)

 # configuration file
 INCLUDE(CheckIncludeFiles)
--- a/src/nvcore/CMakeLists.txt
+++ b/src/nvcore/CMakeLists.txt
@ -19,21 +19,14 @@ SET(CORE_SRCS
 	TextWriter.h
 	TextWriter.cpp
 	Radix.h
-	Radix.cpp
-	Library.h
-	Library.cpp)
+	Radix.cpp)

 INCLUDE_DIRECTORIES(${CMAKE_CURRENT_SOURCE_DIR})

 # targets
 ADD_DEFINITIONS(-DNVCORE_EXPORTS)

-IF(UNIX)
-	SET(LIBS ${LIBS} ${CMAKE_DL_LIBS})
-ENDIF(UNIX)
-
 IF(NVCORE_SHARED)
-	ADD_DEFINITIONS(-DNVCORE_SHARED=1)
 	ADD_LIBRARY(nvcore SHARED ${CORE_SRCS})
 ELSE(NVCORE_SHARED)
 	ADD_LIBRARY(nvcore ${CORE_SRCS})
--- a/src/nvcore/Containers.h
+++ b/src/nvcore/Containers.h
@ -824,13 +824,13 @@ namespace nv
 		}
 	
 		/// Number of entries in the hash.
-		int	size() const
+		int	size()
 		{
 			return entry_count;
 		}
 	
 		/// Number of entries in the hash.
-		int	count() const
+		int	count()
 		{
 			return size();
 		}
--- a/src/nvcore/Debug.cpp
+++ b/src/nvcore/Debug.cpp
@ -28,7 +28,7 @@
 #endif

 #if NV_OS_LINUX && defined(HAVE_EXECINFO_H)
-#	include <execinfo.h> // backtrace
+#	include <execinfo.h>
 #	if NV_CC_GNUC // defined(HAVE_CXXABI_H)
 #		include <cxxabi.h>
 #	endif
@ -38,14 +38,7 @@
 #	include <unistd.h>	// getpid
 #	include <sys/types.h>
 #	include <sys/sysctl.h>	// sysctl
-#	include <sys/ucontext.h>
-#	undef HAVE_EXECINFO_H
-#	if defined(HAVE_EXECINFO_H) // only after OSX 10.5
-#		include <execinfo.h> // backtrace
-#		if NV_CC_GNUC // defined(HAVE_CXXABI_H)
-#			include <cxxabi.h>
-#		endif
-#	endif
+#	include <ucontext.h>
 #endif

 #include <stdexcept> // std::runtime_error
@ -135,14 +128,6 @@ namespace

 #if defined(HAVE_EXECINFO_H) // NV_OS_LINUX

-	static bool nvHasStackTrace() {
-#if NV_OS_DARWIN
-		return backtrace != NULL;
-#else
-		return true;
-#endif
-	}
-
 	static void nvPrintStackTrace(void * trace[], int size, int start=0) {
 		char ** string_array = backtrace_symbols(trace, size);
 	
@ -181,36 +166,24 @@ namespace

 	static void * callerAddress(void * secret)
 	{
-#	if NV_OS_DARWIN
-#		if defined(_STRUCT_MCONTEXT)
-#			if NV_CPU_PPC
-				ucontext_t * ucp = (ucontext_t *)secret;
-				return (void *) ucp->uc_mcontext->__ss.__srr0;
-#			elif NV_CPU_X86
-				ucontext_t * ucp = (ucontext_t *)secret;
-				return (void *) ucp->uc_mcontext->__ss.__eip;
-#			endif
-#		else
-#			if NV_CPU_PPC
-				ucontext_t * ucp = (ucontext_t *)secret;
-				return (void *) ucp->uc_mcontext->ss.srr0;
-#			elif NV_CPU_X86
-				ucontext_t * ucp = (ucontext_t *)secret;
-				return (void *) ucp->uc_mcontext->ss.eip;
-#			endif
-#		endif
+#	if NV_OS_DARWIN && NV_CPU_PPC
+		ucontext_t * ucp = (ucontext_t *)secret;
+        return (void *) ucp->uc_mcontext->ss.srr0;
+#	elif NV_OS_DARWIN && NV_CPU_X86
+		ucontext_t * ucp = (ucontext_t *)secret;
+        return (void *) ucp->uc_mcontext->ss.eip;
+#	elif NV_CPU_X86_64
+		// #define REG_RIP REG_INDEX(rip) // seems to be 16
+		ucontext_t * ucp = (ucontext_t *)secret;
+		return (void *)ucp->uc_mcontext.gregs[REG_RIP];
+#	elif NV_CPU_X86
+		ucontext_t * ucp = (ucontext_t *)secret;
+		return (void *)ucp->uc_mcontext.gregs[14/*REG_EIP*/];
+#	elif NV_CPU_PPC
+		ucontext_t * ucp = (ucontext_t *)secret;
+		return (void *) ucp->uc_mcontext.regs->nip;
 #	else
-#		if NV_CPU_X86_64
-			// #define REG_RIP REG_INDEX(rip) // seems to be 16
-			ucontext_t * ucp = (ucontext_t *)secret;
-			return (void *)ucp->uc_mcontext.gregs[REG_RIP];
-#		elif NV_CPU_X86
-			ucontext_t * ucp = (ucontext_t *)secret;
-			return (void *)ucp->uc_mcontext.gregs[14/*REG_EIP*/];
-#		elif NV_CPU_PPC
-			ucontext_t * ucp = (ucontext_t *)secret;
-			return (void *) ucp->uc_mcontext.regs->nip;
-#		endif
+		return NULL;
 #	endif
 		
 		// How to obtain the instruction pointers in different platforms, from mlton's source code.
@ -255,18 +228,17 @@ namespace
 		}
 		
 #	if defined(HAVE_EXECINFO_H)
-		if (nvHasStackTrace()) // in case of weak linking
-		{
-			void * trace[64];
-			int size = backtrace(trace, 64);
 		
-			if (pnt != NULL) {
-				// Overwrite sigaction with caller's address.
-				trace[1] = pnt;
-			}
-			
-			nvPrintStackTrace(trace, size, 1);
+		void * trace[64];
+		int size = backtrace(trace, 64);
+		
+		if (pnt != NULL) {
+			// Overwrite sigaction with caller's address.
+			trace[1] = pnt;
 		}
+		
+		nvPrintStackTrace(trace, size, 1);
+		
 #	endif // defined(HAVE_EXECINFO_H)
 		
 		exit(0);
@ -401,12 +373,9 @@ namespace
 #		endif

 #		if defined(HAVE_EXECINFO_H)
-			if (nvHasStackTrace())
-			{
-				void * trace[64];
-				int size = backtrace(trace, 64);
-				nvPrintStackTrace(trace, size, 2);
-			}
+			void * trace[64];
+			int size = backtrace(trace, 64);
+			nvPrintStackTrace(trace, size, 3);
 #		endif

 			// Exit cleanly.
@ -453,12 +422,9 @@ void NV_CDECL nvDebug(const char *msg, ...)
 void debug::dumpInfo()
 {
 #if !NV_OS_WIN32 && defined(HAVE_SIGNAL_H) && defined(HAVE_EXECINFO_H)
-	if (nvHasStackTrace())
-	{
-		void * trace[64];
-		int size = backtrace(trace, 64);
-		nvPrintStackTrace(trace, size, 1);
-	}
+	void * trace[64];
+	int size = backtrace(trace, 64);
+	nvPrintStackTrace(trace, size, 1);
 #endif
 }

--- a/src/nvcore/Debug.h
+++ b/src/nvcore/Debug.h
@ -115,7 +115,6 @@ namespace nv
 	{
 		NVCORE_API void dumpInfo();
 	
-		// These functions are not thread safe.
 		NVCORE_API void setMessageHandler( MessageHandler * messageHandler );
 		NVCORE_API void resetMessageHandler();
 	
--- a/src/nvcore/DefsGnucDarwin.h
+++ b/src/nvcore/DefsGnucDarwin.h
@ -2,7 +2,8 @@
 #error "Do not include this file directly."
 #endif

-#include <stdint.h> // uint8_t, int8_t, ...
+#include <stdlib.h>	// uint8_t, int8_t, ...
+

 // Function linkage
 #define DLL_IMPORT
--- a/src/nvcore/DefsVcWin32.h
+++ b/src/nvcore/DefsVcWin32.h
@ -19,9 +19,7 @@

 // Set standard function names.
 #define snprintf _snprintf
-#if _MSC_VER < 1500
-#	define vsnprintf _vsnprintf
-#endif
+#define vsnprintf _vsnprintf
 #define vsscanf _vsscanf
 #define chdir _chdir
 #define getcwd _getcwd 
@ -72,6 +70,8 @@ typedef uint32              uint;
 #pragma warning(disable : 4711)		// function selected for automatic inlining
 #pragma warning(disable : 4725)		// Pentium fdiv bug

+#pragma warning(disable : 4345)		// behavior change: an object of POD type constructed with an initializer of the form () will be default-initialized
+
 #pragma warning(disable : 4786)		// Identifier was truncated and cannot be debugged.

 #pragma warning(disable : 4675)		// resolved overload was found by argument-dependent lookup
--- a/src/nvcore/Library.cpp
+++ b/src/nvcore/Library.cpp
@ -1,41 +0,0 @@
-
-#include "Library.h"
-#include "Debug.h"
-
-#if NV_OS_WIN32
-#define WIN32_LEAN_AND_MEAN
-#define VC_EXTRALEAN
-#include <windows.h>
-#else
-#include <dlfcn.h>
-#endif
-
-
-
-void * nvLoadLibrary(const char * name)
-{
-#if NV_OS_WIN32
-	return (void *)LoadLibraryExA( name, NULL, 0 );
-#else
-	return dlopen(name, RTLD_LAZY);
-#endif
-}
-
-void nvUnloadLibrary(void * handle)
-{
-	nvDebugCheck(handle != NULL);
-#if NV_OS_WIN32
-	FreeLibrary((HMODULE)handle);
-#else
-	dlclose(handle);
-#endif
-}
-
-void * nvBindSymbol(void * handle, const char * symbol)
-{
-#if NV_OS_WIN32
-	return (void *)GetProcAddress((HMODULE)handle, symbol);
-#else
-	return (void *)dlsym(handle, symbol);
-#endif	
-}
--- a/src/nvcore/Library.h
+++ b/src/nvcore/Library.h
@ -1,50 +0,0 @@
-// This code is in the public domain -- castano@gmail.com
-
-#ifndef NV_CORE_LIBRARY_H
-#define NV_CORE_LIBRARY_H
-
-#include <nvcore/nvcore.h>
-
-#if NV_OS_WIN32
-#define LIBRARY_NAME(name)	#name ".dll"
-#elif NV_OS_DARWIN
-#define NV_LIBRARY_NAME(name)	"lib" #name ".dylib"
-#else
-#define NV_LIBRARY_NAME(name)	"lib" #name ".so"
-#endif
-
-NVCORE_API void * nvLoadLibrary(const char * name);
-NVCORE_API void nvUnloadLibrary(void * lib);
-NVCORE_API void * nvBindSymbol(void * lib, const char * symbol);
-
-class NVCORE_CLASS Library
-{
-public:
-	Library(const char * name)
-	{
-		handle = nvLoadLibrary(name);
-	}
-	~Library()
-	{
-		if (isValid())
-		{
-			nvUnloadLibrary(handle);
-		}
-	}
-	
-	bool isValid() const
-	{
-		return handle != NULL;
-	}
-	
-	void * bindSymbol(const char * symbol)
-	{
-		return nvBindSymbol(handle, symbol);
-	}
-	
-private:
-	void * handle;
-};
-
-
-#endif // NV_CORE_LIBRARY_H
--- a/src/nvcore/Prefetch.h
+++ b/src/nvcore/Prefetch.h
@ -24,7 +24,7 @@ __forceinline void nvPrefetch(const void * mem)
 #else // NV_CC_MSVC

 // do nothing in other case.
-#define nvPrefetch(ptr)
+#define piPrefetch(ptr)

 #endif // NV_CC_MSVC

--- a/src/nvcore/Ptr.h
+++ b/src/nvcore/Ptr.h
@ -43,11 +43,8 @@ public:

 	/** Delete owned pointer and assign new one. */
 	void operator=( T * p ) {
-		if (p != m_ptr)
-		{
-			delete m_ptr;
-			m_ptr = p;
-		}
+		delete m_ptr;
+		m_ptr = p;
 	}

 	/** Member access. */
@ -252,14 +249,14 @@ public:
 		/** -> operator. */
 		BaseClass * operator -> () const
 		{
-			nvCheck( m_ptr != NULL );
+			piCheck( m_ptr != NULL );
 			return m_ptr;
 		}

 		/** * operator. */
 		BaseClass & operator*() const
 		{
-			nvCheck( m_ptr != NULL );
+			piCheck( m_ptr != NULL );
 			return *m_ptr;
 		}

--- a/src/nvcore/StdStream.h
+++ b/src/nvcore/StdStream.h
@ -47,25 +47,25 @@ public:

 	/** @name Stream implementation. */
 	//@{
-		virtual void seek( uint pos )
+		virtual void seek( int pos )
 		{
 			nvDebugCheck(m_fp != NULL);
-			nvDebugCheck(pos < size());
+			nvDebugCheck(pos >= 0 && pos < size());
 			fseek(m_fp, pos, SEEK_SET);
 		}
 		
-		virtual uint tell() const
+		virtual int tell() const
 		{
 			nvDebugCheck(m_fp != NULL);
 			return ftell(m_fp);
 		}
 		
-		virtual uint size() const
+		virtual int size() const
 		{
 			nvDebugCheck(m_fp != NULL);
-			uint pos = ftell(m_fp);
+			int pos = ftell(m_fp);
 			fseek(m_fp, 0, SEEK_END);
-			uint end = ftell(m_fp);
+			int end = ftell(m_fp);
 			fseek(m_fp, pos, SEEK_SET);
 			return end;
 		}
@ -117,11 +117,11 @@ public:
 	/** @name Stream implementation. */
 	//@{
 		/// Write data.
-		virtual uint serialize( void * data, uint len )
+		virtual void serialize( void * data, int len )
 		{
 			nvDebugCheck(data != NULL);
 			nvDebugCheck(m_fp != NULL);
-			return (uint)fwrite(data, 1, len, m_fp);
+			fwrite(data, len, 1, m_fp);
 		}
 		
 		virtual bool isLoading() const
@ -156,11 +156,11 @@ public:
 	/** @name Stream implementation. */
 	//@{
 		/// Read data.
-		virtual uint serialize( void * data, uint len )
+		virtual void serialize( void * data, int len )
 		{
 			nvDebugCheck(data != NULL);
 			nvDebugCheck(m_fp != NULL);
-			return (uint)fread(data, 1, len, m_fp);
+			fread(data, len, 1, m_fp);
 		}
 		
 		virtual bool isLoading() const
@ -184,40 +184,33 @@ class NVCORE_CLASS MemoryInputStream : public Stream
 public:

 	/// Ctor.
-	MemoryInputStream( const uint8 * mem, uint size ) : 
+	MemoryInputStream( const uint8 * mem, int size ) : 
 		m_mem(mem), m_ptr(mem), m_size(size) { }

 	/** @name Stream implementation. */
 	//@{
 		/// Read data.
-		virtual uint serialize( void * data, uint len )
+		virtual void serialize( void * data, int len )
 		{
 			nvDebugCheck(data != NULL);
 			nvDebugCheck(!isError());
-
-			uint left = m_size - tell();
-			if (len > left) len = left;
-			
 			memcpy( data, m_ptr, len );
 			m_ptr += len;
-				
-			return len;
 		}
 		
-		virtual void seek( uint pos )
+		virtual void seek( int pos )
 		{
 			nvDebugCheck(!isError());
 			m_ptr = m_mem + pos;
 			nvDebugCheck(!isError());
 		}
 		
-		virtual uint tell() const
+		virtual int tell() const
 		{
-			nvDebugCheck(m_ptr >= m_mem);
-			return uint(m_ptr - m_mem);
+			return int(m_ptr - m_mem);
 		}
 		
-		virtual uint size() const
+		virtual int size() const
 		{
 			return m_size;
 		}
@ -259,7 +252,7 @@ private:

 	const uint8 * m_mem;
 	const uint8 * m_ptr;
-	uint m_size;
+	int m_size;

 };

@ -293,19 +286,17 @@ public:
 	/** @name Stream implementation. */
 	//@{
 		/// Read data.
-		virtual uint serialize( void * data, uint len )
+		virtual void serialize( void * data, int len )
 		{
 			nvDebugCheck(data != NULL);
-			len = m_s->serialize( data, len );
+			m_s->serialize( data, len );
 			
 			if( m_s->isError() ) {
 				throw std::exception();
 			}
-
-			return len;
 		}
 		
-		virtual void seek( uint pos )
+		virtual void seek( int pos )
 		{
 			m_s->seek( pos );
 			
@ -314,12 +305,12 @@ public:
 			}
 		}
 		
-		virtual uint tell() const
+		virtual int tell() const
 		{
 			return m_s->tell();
 		}
 		
-		virtual uint size() const
+		virtual int size() const
 		{
 			return m_s->size();
 		}
--- a/src/nvcore/StrLib.cpp
+++ b/src/nvcore/StrLib.cpp
@ -209,11 +209,48 @@ StringBuilder::StringBuilder( const StringBuilder & s ) : m_size(0), m_str(NULL)
 }

 /** Copy string. */
-StringBuilder::StringBuilder( const char * s ) : m_size(0), m_str(NULL)
+StringBuilder::StringBuilder( const char * s )
 {
 	copy(s);
 }

+/** Allocate and copy string. */
+StringBuilder::StringBuilder( int size_hint, const StringBuilder & s) : m_size(size_hint), m_str(NULL)
+{
+	nvDebugCheck(m_size > 0);
+	m_str = strAlloc(m_size);
+	copy(s);
+}
+
+/** Allocate and format string. */
+StringBuilder::StringBuilder( const char * fmt, ... ) : m_size(0), m_str(NULL)
+{
+	nvDebugCheck(fmt != NULL);
+	va_list arg;
+	va_start( arg, fmt );
+
+	format( fmt, arg );
+
+	va_end( arg );
+}
+
+/** Allocate and format string. */
+StringBuilder::StringBuilder( int size_hint, const char * fmt, ... ) : m_size(size_hint), m_str(NULL)
+{
+	nvDebugCheck(m_size > 0);	
+	nvDebugCheck(fmt != NULL);
+	
+	m_str = strAlloc(m_size);
+
+	va_list arg;
+	va_start( arg, fmt );
+
+	format( fmt, arg );
+
+	va_end( arg );
+}
+
+
 /** Delete the string. */
 StringBuilder::~StringBuilder()
 {
@ -241,7 +278,8 @@ StringBuilder & StringBuilder::format( const char * fmt, ... )
 /** Format a string safely. */
 StringBuilder & StringBuilder::format( const char * fmt, va_list arg )
 {
-	nvDebugCheck(fmt != NULL);
+	nvCheck(fmt != NULL);
+	nvCheck(m_size >= 0);

 	if( m_size == 0 ) {
 		m_size = 64;
@ -289,7 +327,8 @@ StringBuilder & StringBuilder::format( const char * fmt, va_list arg )
 /** Append a string. */
 StringBuilder & StringBuilder::append( const char * s )
 {
-	nvDebugCheck(s != NULL);
+	nvCheck(s != NULL);
+	nvCheck(m_size >= 0);

 	const uint slen = uint(strlen( s ));

@ -436,6 +475,31 @@ void StringBuilder::reset()
 }


+Path::Path(const char * fmt, ...)
+{
+	nvDebugCheck( fmt != NULL );
+
+	va_list arg;
+	va_start( arg, fmt );
+
+	format( fmt, arg );
+
+	va_end( arg );
+}
+
+Path::Path(int size_hint, const char * fmt, ...) : StringBuilder(size_hint)
+{
+	nvDebugCheck( fmt != NULL );
+
+	va_list arg;
+	va_start( arg, fmt );
+
+	format( fmt, arg );
+
+	va_end( arg );
+}
+
+
 /// Get the file name from a path.
 const char * Path::fileName() const
 {
@ -545,6 +609,8 @@ const char * Path::extension(const char * str)
 }


+// static
+String String::s_null(String::null);

 /// Clone this string
 String String::clone() const
@ -555,13 +621,13 @@ String String::clone() const

 void String::setString(const char * str)
 {
-	if (str == NULL) {
-		data = NULL;
+	if( str == NULL ) {
+		data = s_null.data;
 	}
 	else {
 		allocString( str );
-		addRef();
 	}
+	addRef();
 }

 void String::setString(const char * str, int length)
@ -574,11 +640,11 @@ void String::setString(const char * str, int length)

 void String::setString(const StringBuilder & str)
 {
-	if (str.str() == NULL) {
-		data =	NULL;
+	if( str.str() == NULL ) {
+		data = s_null.data;
 	}
 	else {
 		allocString(str);
-		addRef();
 	}
+	addRef();
 }	
--- a/src/nvcore/StrLib.h
+++ b/src/nvcore/StrLib.h
@ -14,10 +14,10 @@ namespace nv

 	uint strHash(const char * str, uint h) NV_PURE;

-	/// String hash based on Bernstein's hash.
+	/// String hash vased on Bernstein's hash.
 	inline uint strHash(const char * data, uint h = 5381)
 	{
-		uint i = 0;
+		uint i;
 		while(data[i] != 0) {
 			h = (33 * h) ^ uint(data[i]);
 			i++;
@ -47,6 +47,9 @@ namespace nv
 		explicit StringBuilder( int size_hint );
 		StringBuilder( const char * str );
 		StringBuilder( const StringBuilder & );
+		StringBuilder( int size_hint, const StringBuilder & );	
+		StringBuilder( const char * format, ... ) __attribute__((format (printf, 2, 3)));
+		StringBuilder( int size_hint, const char * format, ... ) __attribute__((format (printf, 3, 4)));
 	
 		~StringBuilder();
 	
@ -117,16 +120,18 @@ namespace nv
 		char * m_str;
 		
 	};
+	

-
-	/// Path string. @@ This should be called PathBuilder.
+	/// Path string.
 	class NVCORE_CLASS Path : public StringBuilder
 	{
 	public:
 		Path() : StringBuilder() {}
 		explicit Path(int size_hint) : StringBuilder(size_hint) {}
-		Path(const char * str) : StringBuilder(str) {}
-		Path(const Path & path) : StringBuilder(path) {}
+		Path(const StringBuilder & str) : StringBuilder(str) {}
+		Path(int size_hint, const StringBuilder & str) : StringBuilder(size_hint, str) {}	
+		Path(const char * format, ...) __attribute__((format (printf, 2, 3)));
+		Path(int size_hint, const char * format, ...) __attribute__((format (printf, 3, 4)));
 		
 		const char * fileName() const;
 		const char * extension() const;
@ -135,7 +140,7 @@ namespace nv
 		
 		void stripFileName();
 		void stripExtension();
-
+		
 		// statics
 		NVCORE_API static char separator();
 		NVCORE_API static const char * fileName(const char *);
@ -151,14 +156,15 @@ namespace nv
 		/// Constructs a null string. @sa isNull()
 		String()
 		{
-			data = NULL;
+			data = s_null.data;
+			addRef();
 		}

 		/// Constructs a shared copy of str.
 		String(const String & str)
 		{
 			data = str.data;
-			if (data != NULL) addRef();
+			addRef();
 		}

 		/// Constructs a shared string from a standard string.
@ -182,6 +188,7 @@ namespace nv
 		/// Dtor.
 		~String()
 		{
+			nvDebugCheck(data != NULL);
 			release();
 		}

@ -206,61 +213,52 @@ namespace nv
 		/// Implement value semantics.
 		String & operator=( const String & str )
 		{
-			if (str.data != data)
-			{
-				release();
-				data = str.data;
-				addRef();
-			}
+			release();
+			data = str.data;
+			addRef();
 			return *this;
 		}

 		/// Equal operator.
 		bool operator==( const String & str ) const
 		{
+			nvDebugCheck(data != NULL);
+			nvDebugCheck(str.data != NULL);
 			if( str.data == data ) {
 				return true;
 			}
-			if ((data == NULL) != (str.data == NULL)) {
-				return false;
-			}
 			return strcmp(data, str.data) == 0;
 		}

 		/// Equal operator.
 		bool operator==( const char * str ) const
 		{
+			nvDebugCheck(data != NULL);
 			nvCheck(str != NULL);	// Use isNull!
-			if (data == NULL) {
-				return false;
-			}
 			return strcmp(data, str) == 0;
 		}

 		/// Not equal operator.
 		bool operator!=( const String & str ) const
 		{
+			nvDebugCheck(data != NULL);
+			nvDebugCheck(str.data != NULL);
 			if( str.data == data ) {
 				return false;
 			}
-			if ((data == NULL) != (str.data == NULL)) {
-				return true;
-			}
 			return strcmp(data, str.data) != 0;
 		}
 	
 		/// Not equal operator.
 		bool operator!=( const char * str ) const
 		{
+			nvDebugCheck(data != NULL);
 			nvCheck(str != NULL);	// Use isNull!
-			if (data == NULL) {
-				return false;
-			}
 			return strcmp(data, str) != 0;
 		}
 	
 		/// Returns true if this string is the null string.
-		bool isNull() const { return data == NULL; }
+		bool isNull() const { nvDebugCheck(data != NULL); return data == s_null.data; }
 	
 		/// Return the exact length.
 		uint length() const { nvDebugCheck(data != NULL); return uint(strlen(data)); }
@ -269,45 +267,44 @@ namespace nv
 		uint hash() const { nvDebugCheck(data != NULL); return strHash(data); }
 	
 		/// const char * cast operator.
-		operator const char * () const { return data; }
+		operator const char * () const { nvDebugCheck(data != NULL); return data; }
 	
 		/// Get string pointer.
-		const char * str() const { return data; }
+		const char * str() const { nvDebugCheck(data != NULL); return data; }
 	

 	private:

+		enum null_t { null };
+		
+		// Private constructor for null string.
+		String(null_t) {
+			setString("");
+		}
+
 		// Add reference count.
-		void addRef()
-		{
-			if (data != NULL)
-			{
-				setRefCount(getRefCount() + 1);
-			}
+		void addRef() {
+			nvDebugCheck(data != NULL);
+			setRefCount(getRefCount() + 1);
 		}
 		
 		// Decrease reference count.
-		void release()
-		{
-			if (data != NULL)
-			{
-				const uint16 count = getRefCount();
-				setRefCount(count - 1);
-				if (count - 1 == 0) {
-					mem::free(data - 2);
-					data = NULL;
-				}
+		void release() {
+			nvDebugCheck(data != NULL);
+
+			const uint16 count = getRefCount();
+			setRefCount(count - 1);
+			if( count - 1 == 0 ) {
+				mem::free(data - 2);
+				data = NULL;
 			}
 		}
 		
-		uint16 getRefCount() const
-		{
-			nvDebugCheck(data != NULL);
+		uint16 getRefCount() const {
 			return *reinterpret_cast<const uint16 *>(data - 2);
 		}
 		
 		void setRefCount(uint16 count) {
-			nvDebugCheck(data != NULL);
 			nvCheck(count < 0xFFFF);
 			*reinterpret_cast<uint16 *>(const_cast<char *>(data - 2)) = uint16(count);
 		}
@ -346,6 +343,8 @@ namespace nv
 	
 	private:

+		NVCORE_API static String s_null;
+
 		const char * data;
 		
 	};
--- a/src/nvcore/Stream.h
+++ b/src/nvcore/Stream.h
@ -41,17 +41,17 @@ public:
 	ByteOrder byteOrder() const { return m_byteOrder; }

 	
-	/// Serialize the given data.
-	virtual uint serialize( void * data, uint len ) = 0;
+	/// Serialize the given data. @@ Should return bytes serialized?
+	virtual void serialize( void * data, int len ) = 0;

 	/// Move to the given position in the archive.
-	virtual void seek( uint pos ) = 0;
+	virtual void seek( int pos ) = 0;

 	/// Return the current position in the archive.
-	virtual uint tell() const = 0;
+	virtual int tell() const = 0;

 	/// Return the current size of the archive.
-	virtual uint size() const = 0;
+	virtual int size() const = 0;

 	/// Determine if there has been any error.
 	virtual bool isError() const = 0;
@ -136,13 +136,13 @@ public:
 protected:

 	/// Serialize in the stream byte order.
-	Stream & byteOrderSerialize( void * v, uint len ) {
+	Stream & byteOrderSerialize( void * v, int len ) {
 		if( m_byteOrder == getSystemByteOrder() ) {
 			serialize( v, len );
 		}
 		else {
-			for( uint i = len; i > 0; i-- ) {
-				serialize( (uint8 *)v + i - 1, 1 );
+			for( int i=len-1; i>=0; i-- ) {
+				serialize( (uint8 *)v + i, 1 );
 			}
 		}
 		return *this;
--- a/src/nvimage/CMakeLists.txt
+++ b/src/nvimage/CMakeLists.txt
@ -53,7 +53,6 @@ ENDIF(OPENEXR_FOUND)
 ADD_DEFINITIONS(-DNVIMAGE_EXPORTS)

 IF(NVIMAGE_SHARED)	
-	ADD_DEFINITIONS(-DNVIMAGE_SHARED=1)
 	ADD_LIBRARY(nvimage SHARED ${IMAGE_SRCS})
 ELSE(NVIMAGE_SHARED)
 	ADD_LIBRARY(nvimage ${IMAGE_SRCS})
--- a/src/nvimage/ColorBlock.cpp
+++ b/src/nvimage/ColorBlock.cpp
@ -113,18 +113,15 @@ void ColorBlock::splatY()
 /// Returns true if the block has a single color.
 bool ColorBlock::isSingleColor() const
 {
-	Color32 mask(0xFF, 0xFF, 0xFF, 0x00);
-	uint u = m_color[0].u & mask.u;
-        
-	for (int i = 1; i < 16; i++)
-	{
-		if (u != (m_color[i].u & mask.u))
-		{
-			return false;
-		}
-	}
-        
-	return true;
+	for(int i = 1; i < 16; i++)
+	{
+		if (m_color[0] != m_color[i])
+		{
+			return false;
+		}
+	}
+	
+	return true;
 }

 /// Count number of unique colors in this color block.
--- a/src/nvimage/DirectDrawSurface.cpp
+++ b/src/nvimage/DirectDrawSurface.cpp
@ -532,7 +532,7 @@ DDSHeader::DDSHeader()

 	// Store version information on the reserved header attributes.
 	this->reserved[9] = MAKEFOURCC('N', 'V', 'T', 'T');
-	this->reserved[10] = (2 << 16) | (0 << 8) | (8);	// major.minor.revision
+	this->reserved[10] = (2 << 16) | (0 << 8) | (4);	// major.minor.revision

 	this->pf.size = 32;
 	this->pf.flags = 0;
@ -989,10 +989,10 @@ void DirectDrawSurface::readLinearImage(Image * img)
 			stream->serialize(&c, byteCount);

 			Color32 pixel(0, 0, 0, 0xFF);
-			pixel.r = PixelFormat::convert((c & header.pf.rmask) >> rshift, rsize, 8);
-			pixel.g = PixelFormat::convert((c & header.pf.gmask) >> gshift, gsize, 8);
-			pixel.b = PixelFormat::convert((c & header.pf.bmask) >> bshift, bsize, 8);
-			pixel.a = PixelFormat::convert((c & header.pf.amask) >> ashift, asize, 8);
+			pixel.r = PixelFormat::convert(c >> rshift, rsize, 8);
+			pixel.g = PixelFormat::convert(c >> gshift, gsize, 8);
+			pixel.b = PixelFormat::convert(c >> bshift, bsize, 8);
+			pixel.a = PixelFormat::convert(c >> ashift, asize, 8);

 			img->pixel(x, y) = pixel;
 		}
--- a/src/nvimage/Filter.cpp
+++ b/src/nvimage/Filter.cpp
@ -33,10 +33,11 @@
 * http://www.dspguide.com/ch16.htm
 */

-#include "Filter.h"

-#include <nvmath/Vector.h>	// Vector4
 #include <nvcore/Containers.h>	// swap
+#include <nvmath/nvmath.h>	// fabs
+#include <nvmath/Vector.h>	// Vector4
+#include <nvimage/Filter.h>

 using namespace nv;

@ -503,7 +504,7 @@ void Kernel2::initBlendedSobel(const Vector4 & scale)

 		for (int i = 0; i < 7; i++) {
 			for (int e = 0; e < 7; e++) {
-				m_data[(i + 1) * 9 + e + 1] += elements[i * 7 + e] * scale.z();
+				m_data[i * 9 + e + 1] += elements[i * 7 + e] * scale.z();
 			}
 		}
 	}
@ -518,7 +519,7 @@ void Kernel2::initBlendedSobel(const Vector4 & scale)

 		for (int i = 0; i < 5; i++) {
 			for (int e = 0; e < 5; e++) {
-				m_data[(i + 2) * 9 + e + 2] += elements[i * 5 + e] * scale.y();
+				m_data[i * 9 + e + 2] += elements[i * 5 + e] * scale.y();
 			}
 		}
 	}
@ -531,7 +532,7 @@ void Kernel2::initBlendedSobel(const Vector4 & scale)

 		for (int i = 0; i < 3; i++) {
 			for (int e = 0; e < 3; e++) {
-				m_data[(i + 3) * 9 + e + 3] += elements[i * 3 + e] * scale.x();
+				m_data[i * 9 + e + 3] += elements[i * 3 + e] * scale.x();
 			}
 		}
 	}
@ -581,6 +582,7 @@ PolyphaseKernel::PolyphaseKernel(const Filter & f, uint srcLength, uint dstLengt
 			m_data[i * m_windowSize + j] /= total;
 		}
 	}
+
 }

 PolyphaseKernel::~PolyphaseKernel()
--- a/src/nvimage/Filter.h
+++ b/src/nvimage/Filter.h
@ -11,16 +11,16 @@ namespace nv
 	class Vector4;

 	/// Base filter class.
-	class NVIMAGE_CLASS Filter
+	class Filter
 	{
 	public:
-		Filter(float width);
-		virtual ~Filter();
+		NVIMAGE_API Filter(float width);
+		NVIMAGE_API virtual ~Filter();

-		float width() const { return m_width; }
-		float sampleDelta(float x, float scale) const;
-		float sampleBox(float x, float scale, int samples) const;
-		float sampleTriangle(float x, float scale, int samples) const;
+		NVIMAGE_API float width() const { return m_width; }
+		NVIMAGE_API float sampleDelta(float x, float scale) const;
+		NVIMAGE_API float sampleBox(float x, float scale, int samples) const;
+		NVIMAGE_API float sampleTriangle(float x, float scale, int samples) const;

 		virtual float evaluate(float x) const = 0;

@ -29,56 +29,56 @@ namespace nv
 	};

 	// Box filter.
-	class NVIMAGE_CLASS BoxFilter : public Filter
+	class BoxFilter : public Filter
 	{
 	public:
-		BoxFilter();
-		BoxFilter(float width);
-		virtual float evaluate(float x) const;
+		NVIMAGE_API BoxFilter();
+		NVIMAGE_API BoxFilter(float width);
+		NVIMAGE_API virtual float evaluate(float x) const;
 	};

 	// Triangle (bilinear/tent) filter.
-	class NVIMAGE_CLASS TriangleFilter : public Filter
+	class TriangleFilter : public Filter
 	{
 	public:
-		TriangleFilter();
-		TriangleFilter(float width);
-		virtual float evaluate(float x) const;
+		NVIMAGE_API TriangleFilter();
+		NVIMAGE_API TriangleFilter(float width);
+		NVIMAGE_API virtual float evaluate(float x) const;
 	};

 	// Quadratic (bell) filter.
-	class NVIMAGE_CLASS QuadraticFilter : public Filter
+	class QuadraticFilter : public Filter
 	{
 	public:
-		QuadraticFilter();
-		virtual float evaluate(float x) const;
+		NVIMAGE_API QuadraticFilter();
+		NVIMAGE_API virtual float evaluate(float x) const;
 	};

 	// Cubic filter from Thatcher Ulrich.
-	class NVIMAGE_CLASS CubicFilter : public Filter
+	class CubicFilter : public Filter
 	{
 	public:
-		CubicFilter();
-		virtual float evaluate(float x) const;
+		NVIMAGE_API CubicFilter();
+		NVIMAGE_API virtual float evaluate(float x) const;
 	};

 	// Cubic b-spline filter from Paul Heckbert.
-	class NVIMAGE_CLASS BSplineFilter : public Filter
+	class BSplineFilter : public Filter
 	{
 	public:
-		BSplineFilter();
-		virtual float evaluate(float x) const;
+		NVIMAGE_API BSplineFilter();
+		NVIMAGE_API virtual float evaluate(float x) const;
 	};

 	/// Mitchell & Netravali's two-param cubic
 	/// @see "Reconstruction Filters in Computer Graphics", SIGGRAPH 88
-	class NVIMAGE_CLASS MitchellFilter : public Filter
+	class MitchellFilter : public Filter
 	{
 	public:
-		MitchellFilter();
-		virtual float evaluate(float x) const;
+		NVIMAGE_API MitchellFilter();
+		NVIMAGE_API virtual float evaluate(float x) const;

-		void setParameters(float b, float c);
+		NVIMAGE_API void setParameters(float a, float b);

 	private:
 		float p0, p2, p3;
@ -86,29 +86,29 @@ namespace nv
 	};

 	// Lanczos3 filter.
-	class NVIMAGE_CLASS LanczosFilter : public Filter
+	class LanczosFilter : public Filter
 	{
 	public:
-		LanczosFilter();
-		virtual float evaluate(float x) const;
+		NVIMAGE_API LanczosFilter();
+		NVIMAGE_API virtual float evaluate(float x) const;
 	};

 	// Sinc filter.
-	class NVIMAGE_CLASS SincFilter : public Filter
+	class SincFilter : public Filter
 	{
 	public:
-		SincFilter(float w);
-		virtual float evaluate(float x) const;
+		NVIMAGE_API SincFilter(float w);
+		NVIMAGE_API virtual float evaluate(float x) const;
 	};

 	// Kaiser filter.
-	class NVIMAGE_CLASS KaiserFilter : public Filter
+	class KaiserFilter : public Filter
 	{
 	public:
-		KaiserFilter(float w);
-		virtual float evaluate(float x) const;
+		NVIMAGE_API KaiserFilter(float w);
+		NVIMAGE_API virtual float evaluate(float x) const;
 	
-		void setParameters(float a, float stretch);
+		NVIMAGE_API void setParameters(float a, float stretch);

 	private:
 		float alpha;
@ -118,12 +118,12 @@ namespace nv


 	/// A 1D kernel. Used to precompute filter weights.
-	class NVIMAGE_CLASS Kernel1
+	class Kernel1
 	{
 		NV_FORBID_COPY(Kernel1);
 	public:
-		Kernel1(const Filter & f, int iscale, int samples = 32);
-		~Kernel1();
+		NVIMAGE_API Kernel1(const Filter & f, int iscale, int samples = 32);
+		NVIMAGE_API ~Kernel1();
 		
 		float valueAt(uint x) const {
 			nvDebugCheck(x < (uint)m_windowSize);
@ -138,7 +138,7 @@ namespace nv
 			return m_width;
 		}
 		
-		void debugPrint();
+		NVIMAGE_API void debugPrint();
 		
 	private:
 		int m_windowSize;
@ -148,15 +148,15 @@ namespace nv


 	/// A 2D kernel.
-	class NVIMAGE_CLASS Kernel2 
+	class Kernel2 
 	{
 	public:
-		Kernel2(uint width);
-		Kernel2(const Kernel2 & k);
-		~Kernel2();
+		NVIMAGE_API Kernel2(uint width);
+		NVIMAGE_API Kernel2(const Kernel2 & k);
+		NVIMAGE_API ~Kernel2();
 		
-		void normalize();
-		void transpose();
+		NVIMAGE_API void normalize();
+		NVIMAGE_API void transpose();
 		
 		float valueAt(uint x, uint y) const {
 			return m_data[y * m_windowSize + x];
@ -166,12 +166,12 @@ namespace nv
 			return m_windowSize;
 		}
 		
-		void initLaplacian();
-		void initEdgeDetection();
-		void initSobel();
-		void initPrewitt();
+		NVIMAGE_API void initLaplacian();
+		NVIMAGE_API void initEdgeDetection();
+		NVIMAGE_API void initSobel();
+		NVIMAGE_API void initPrewitt();
 		
-		void initBlendedSobel(const Vector4 & scale);
+		NVIMAGE_API void initBlendedSobel(const Vector4 & scale);
 		
 	private:
 		const uint m_windowSize;
@ -180,12 +180,12 @@ namespace nv


 	/// A 1D polyphase kernel
-	class NVIMAGE_CLASS PolyphaseKernel
+	class PolyphaseKernel
 	{
 		NV_FORBID_COPY(PolyphaseKernel);
 	public:
-		PolyphaseKernel(const Filter & f, uint srcLength, uint dstLength, int samples = 32);
-		~PolyphaseKernel();
+		NVIMAGE_API PolyphaseKernel(const Filter & f, uint srcLength, uint dstLength, int samples = 32);
+		NVIMAGE_API ~PolyphaseKernel();
 		
 		int windowSize() const {
 			return m_windowSize;
@ -205,7 +205,7 @@ namespace nv
 			return m_data[column * m_windowSize + x];
 		}

-		void debugPrint() const;
+		NVIMAGE_API void debugPrint() const;
 		
 	private:
 		int m_windowSize;
--- a/src/nvimage/FloatImage.cpp
+++ b/src/nvimage/FloatImage.cpp
@ -1,18 +1,16 @@
 // This code is in the public domain -- castanyo@yahoo.es

+#include <nvcore/Containers.h>
+#include <nvcore/Ptr.h>
+
+#include <nvmath/Color.h>
+
 #include "FloatImage.h"
 #include "Filter.h"
 #include "Image.h"

-#include <nvmath/Color.h>
-#include <nvmath/Matrix.h>
-
-#include <nvcore/Containers.h>
-#include <nvcore/Ptr.h>
-
 #include <math.h>

-
 using namespace nv;

 namespace 
@ -142,8 +140,7 @@ Image * FloatImage::createImageGammaCorrect(float gamma/*= 2.2f*/) const
 /// Allocate a 2d float image of the given format and the given extents.
 void FloatImage::allocate(uint c, uint w, uint h)
 {
-	free();
-
+	nvCheck(m_mem == NULL);
 	m_width = w;
 	m_height = h;
 	m_componentNum = c;
@ -154,6 +151,7 @@ void FloatImage::allocate(uint c, uint w, uint h)
 /// Free the image, but don't clear the members.
 void FloatImage::free()
 {
+	nvCheck(m_mem != NULL);
 	nv::mem::free( reinterpret_cast<void *>(m_mem) );
 	m_mem = NULL;
 }
@ -551,15 +549,6 @@ FloatImage * FloatImage::downSample(const Filter & filter, WrapMode wm) const
 	return resize(filter, w, h, wm);
 }

-/// Downsample applying a 1D kernel separately in each dimension.
-FloatImage * FloatImage::downSample(const Filter & filter, WrapMode wm, uint alpha) const
-{
-	const uint w = max(1, m_width / 2);
-	const uint h = max(1, m_height / 2);
-
-	return resize(filter, w, h, wm, alpha);
-}
-

 /// Downsample applying a 1D kernel separately in each dimension.
 FloatImage * FloatImage::resize(const Filter & filter, uint w, uint h, WrapMode wm) const
@ -631,56 +620,10 @@ FloatImage * FloatImage::resize(const Filter & filter, uint w, uint h, WrapMode
 	return dst_image.release();
 }

-/// Downsample applying a 1D kernel separately in each dimension.
-FloatImage * FloatImage::resize(const Filter & filter, uint w, uint h, WrapMode wm, uint alpha) const
-{
-	nvCheck(alpha < m_componentNum);
-
-	AutoPtr<FloatImage> tmp_image( new FloatImage() );
-	AutoPtr<FloatImage> dst_image( new FloatImage() );	
-	
-	PolyphaseKernel xkernel(filter, m_width, w, 32);
-	PolyphaseKernel ykernel(filter, m_height, h, 32);
-	
-	{
-		tmp_image->allocate(m_componentNum, w, m_height);
-		dst_image->allocate(m_componentNum, w, h);
-		
-		Array<float> tmp_column(h);
-		tmp_column.resize(h);
-		
-		for (uint c = 0; c < m_componentNum; c++)
-		{
-			float * tmp_channel = tmp_image->channel(c);
-			
-			for (uint y = 0; y < m_height; y++) {
-				this->applyKernelHorizontal(xkernel, y, c, alpha, wm, tmp_channel + y * w);
-			}
-		}
-
-		// Process all channels before applying vertical kernel to make sure alpha has been computed.
-
-		for (uint c = 0; c < m_componentNum; c++)
-		{
-			float * dst_channel = dst_image->channel(c);
-			
-			for (uint x = 0; x < w; x++) {
-				tmp_image->applyKernelVertical(ykernel, x, c, alpha, wm, tmp_column.unsecureBuffer());
-				
-				for (uint y = 0; y < h; y++) {
-					dst_channel[y * w + x] = tmp_column[y];
-				}
-			}
-		}
-	}
-	
-	return dst_image.release();
-}
-


 /// Apply 2D kernel at the given coordinates and return result.
-float FloatImage::applyKernel(const Kernel2 * k, int x, int y, uint c, WrapMode wm) const
+float FloatImage::applyKernel(const Kernel2 * k, int x, int y, int c, WrapMode wm) const
 {
 	nvDebugCheck(k != NULL);
 	
@ -709,7 +652,7 @@ float FloatImage::applyKernel(const Kernel2 * k, int x, int y, uint c, WrapMode


 /// Apply 1D vertical kernel at the given coordinates and return result.
-float FloatImage::applyKernelVertical(const Kernel1 * k, int x, int y, uint c, WrapMode wm) const
+float FloatImage::applyKernelVertical(const Kernel1 * k, int x, int y, int c, WrapMode wm) const
 {
 	nvDebugCheck(k != NULL);
 	
@ -731,7 +674,7 @@ float FloatImage::applyKernelVertical(const Kernel1 * k, int x, int y, uint c, W
 }

 /// Apply 1D horizontal kernel at the given coordinates and return result.
-float FloatImage::applyKernelHorizontal(const Kernel1 * k, int x, int y, uint c, WrapMode wm) const
+float FloatImage::applyKernelHorizontal(const Kernel1 * k, int x, int y, int c, WrapMode wm) const
 {
 	nvDebugCheck(k != NULL);
 	
@ -754,7 +697,7 @@ float FloatImage::applyKernelHorizontal(const Kernel1 * k, int x, int y, uint c,


 /// Apply 1D vertical kernel at the given coordinates and return result.
-void FloatImage::applyKernelVertical(const PolyphaseKernel & k, int x, uint c, WrapMode wm, float * __restrict output) const
+void FloatImage::applyKernelVertical(const PolyphaseKernel & k, int x, int c, WrapMode wm, float * output) const
 {
 	const uint length = k.length();
 	const float scale = float(length) / float(m_height);
@ -786,7 +729,7 @@ void FloatImage::applyKernelVertical(const PolyphaseKernel & k, int x, uint c, W
 }

 /// Apply 1D horizontal kernel at the given coordinates and return result.
-void FloatImage::applyKernelHorizontal(const PolyphaseKernel & k, int y, uint c, WrapMode wm, float * __restrict output) const
+void FloatImage::applyKernelHorizontal(const PolyphaseKernel & k, int y, int c, WrapMode wm, float * output) const
 {
 	const uint length = k.length();
 	const float scale = float(length) / float(m_width);
@ -817,93 +760,3 @@ void FloatImage::applyKernelHorizontal(const PolyphaseKernel & k, int y, uint c,
 	}
 }

-
-/// Apply 1D vertical kernel at the given coordinates and return result.
-void FloatImage::applyKernelVertical(const PolyphaseKernel & k, int x, uint c, uint a, WrapMode wm, float * __restrict output) const
-{
-	const uint length = k.length();
-	const float scale = float(length) / float(m_height);
-	const float iscale = 1.0f / scale;
-
-	const float width = k.width();
-	const int windowSize = k.windowSize();
-
-	const float * channel = this->channel(c);
-	const float * alpha = this->channel(a);
-
-	for (uint i = 0; i < length; i++)
-	{
-		const float center = (0.5f + i) * iscale;
-		
-		const int left = (int)floorf(center - width);
-		const int right = (int)ceilf(center + width);
-		nvCheck(right - left <= windowSize);
-		
-		float norm = 0;
-		float sum = 0;
-		for (int j = 0; j < windowSize; ++j)
-		{
-			const int idx = this->index(x, j+left, wm);
-			
-			float w = k.valueAt(i, j) * (alpha[idx] + (1.0f / 256.0f));
-			norm += w;
-			sum += w * channel[idx];
-		}
-		
-		output[i] = sum / norm;
-	}
-}
-
-/// Apply 1D horizontal kernel at the given coordinates and return result.
-void FloatImage::applyKernelHorizontal(const PolyphaseKernel & k, int y, uint c, uint a, WrapMode wm, float * __restrict output) const
-{
-	const uint length = k.length();
-	const float scale = float(length) / float(m_width);
-	const float iscale = 1.0f / scale;
-
-	const float width = k.width();
-	const int windowSize = k.windowSize();
-
-	const float * channel = this->channel(c);
-	const float * alpha = this->channel(a);
-
-	for (uint i = 0; i < length; i++)
-	{
-		const float center = (0.5f + i) * iscale;
-		
-		const int left = (int)floorf(center - width);
-		const int right = (int)ceilf(center + width);
-		nvDebugCheck(right - left <= windowSize);
-		
-		float norm = 0.0f;
-		float sum = 0;
-		for (int j = 0; j < windowSize; ++j)
-		{
-			const int idx = this->index(left + j, y, wm);
-
-			float w = k.valueAt(i, j) * (alpha[idx] + (1.0f / 256.0f));
-			norm += w;
-			sum += w * channel[idx];
-		}
-		
-		output[i] = sum / norm;
-	}
-}
-
-FloatImage* FloatImage::clone() const
-{
-	FloatImage* copy = new FloatImage();
-	copy->m_width = m_width;
-	copy->m_height = m_height;
-	copy->m_componentNum = m_componentNum;
-	copy->m_count = m_count;
-	
-	if(m_mem)
-	{
-		copy->allocate(m_componentNum, m_width, m_height);
-		memcpy(copy->m_mem, m_mem, m_count * sizeof(float));
-	}
-		
-	return copy;
-}
-
--- a/src/nvimage/FloatImage.h
+++ b/src/nvimage/FloatImage.h
@ -3,20 +3,12 @@
 #ifndef NV_IMAGE_FLOATIMAGE_H
 #define NV_IMAGE_FLOATIMAGE_H

-#include <nvimage/nvimage.h>
-
-#include <nvmath/Vector.h>
-
 #include <nvcore/Debug.h>
 #include <nvcore/Containers.h> // clamp
-
-#include <stdlib.h> // abs
-
+#include <nvimage/nvimage.h>

 namespace nv
 {
-class Vector4;
-class Matrix;
 class Image;
 class Filter;
 class Kernel1;
@ -68,22 +60,20 @@ public:
 	NVIMAGE_API void toGamma(uint base_component, uint num, float gamma = 2.2f);
 	NVIMAGE_API void exponentiate(uint base_component, uint num, float power);
 	
-
+	
 	NVIMAGE_API FloatImage * fastDownSample() const;
 	NVIMAGE_API FloatImage * downSample(const Filter & filter, WrapMode wm) const;
-	NVIMAGE_API FloatImage * downSample(const Filter & filter, WrapMode wm, uint alpha) const;
 	NVIMAGE_API FloatImage * resize(const Filter & filter, uint w, uint h, WrapMode wm) const;

-	NVIMAGE_API FloatImage * resize(const Filter & filter, uint w, uint h, WrapMode wm, uint alpha) const;
+	//NVIMAGE_API FloatImage * downSample(const Kernel1 & filter, WrapMode wm) const;
+	//NVIMAGE_API FloatImage * downSample(const Kernel1 & filter, uint w, uint h, WrapMode wm) const;
 	//@}

-	NVIMAGE_API float applyKernel(const Kernel2 * k, int x, int y, uint c, WrapMode wm) const;
-	NVIMAGE_API float applyKernelVertical(const Kernel1 * k, int x, int y, uint c, WrapMode wm) const;
-	NVIMAGE_API float applyKernelHorizontal(const Kernel1 * k, int x, int y, uint c, WrapMode wm) const;
-	NVIMAGE_API void applyKernelVertical(const PolyphaseKernel & k, int x, uint c, WrapMode wm, float * output) const;
-	NVIMAGE_API void applyKernelHorizontal(const PolyphaseKernel & k, int y, uint c, WrapMode wm, float * output) const;
-	NVIMAGE_API void applyKernelVertical(const PolyphaseKernel & k, int x, uint c, uint a, WrapMode wm, float * output) const;
-	NVIMAGE_API void applyKernelHorizontal(const PolyphaseKernel & k, int y, uint c, uint a, WrapMode wm, float * output) const;
+	NVIMAGE_API float applyKernel(const Kernel2 * k, int x, int y, int c, WrapMode wm) const;
+	NVIMAGE_API float applyKernelVertical(const Kernel1 * k, int x, int y, int c, WrapMode wm) const;
+	NVIMAGE_API float applyKernelHorizontal(const Kernel1 * k, int x, int y, int c, WrapMode wm) const;
+	NVIMAGE_API void applyKernelVertical(const PolyphaseKernel & k, int x, int c, WrapMode wm, float * output) const;
+	NVIMAGE_API void applyKernelHorizontal(const PolyphaseKernel & k, int y, int c, WrapMode wm, float * output) const;
 	
 	
 	uint width() const { return m_width; }
@ -119,9 +109,6 @@ public:
 	float sampleLinearMirror(float x, float y, int c) const;
 	//@}
 	
-	
-	FloatImage* clone() const;
-	
 public:
 	
 	uint index(uint x, uint y) const;
@ -247,7 +234,7 @@ inline uint FloatImage::indexMirror(int x, int y) const
 	}

 	if (m_height == 1) y = 0;
-
+	
 	y = abs(y);
 	while (y >= m_height) {
 		y = abs(m_height + m_height - y - 2);
--- a/src/nvimage/Image.cpp
+++ b/src/nvimage/Image.cpp
@ -78,7 +78,7 @@ void Image::unwrap()

 void Image::free()
 {
-	nv::mem::free(m_data);
+	::free(m_data);
 	m_data = NULL;
 }

--- a/src/nvimage/NormalMap.cpp
+++ b/src/nvimage/NormalMap.cpp
@ -21,16 +21,15 @@
 // FROM, OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR
 // OTHER DEALINGS IN THE SOFTWARE.

+#include <nvcore/Ptr.h>
+
+#include <nvmath/Color.h>
+
 #include <nvimage/NormalMap.h>
 #include <nvimage/Filter.h>
 #include <nvimage/FloatImage.h>
 #include <nvimage/Image.h>

-#include <nvmath/Color.h>
-
-#include <nvcore/Ptr.h>
-
-
 using namespace nv;

 // Create normal map using the given kernels.
--- a/src/nvimage/PsdFile.h
+++ b/src/nvimage/PsdFile.h
@ -39,7 +39,7 @@ namespace nv
 		bool isSupported() const
 		{
 			if (version != 1) {
-				nvDebug("*** bad version number %u\n", version);
+				printf("*** bad version number %u\n", version);
 				return false;
 			}
 			if (channel_count > 4) {
--- a/src/nvimage/Quantize.cpp
+++ b/src/nvimage/Quantize.cpp
@ -12,14 +12,11 @@ http://www.efg2.com/Lab/Library/ImageProcessing/DHALF.TXT
@@ This code needs to be reviewed, I'm not sure it's correct.
 */

-#include <nvimage/Quantize.h>
-#include <nvimage/Image.h>
-#include <nvimage/PixelFormat.h>
-
 #include <nvmath/Color.h>

-#include <nvcore/Containers.h> // swap
-
+#include <nvimage/Image.h>
+#include <nvimage/Quantize.h>
+#include <nvimage/PixelFormat.h>

 using namespace nv;

@ -136,17 +133,17 @@ void nv::Quantize::Truncate(Image * image, uint rsize, uint gsize, uint bsize, u
 			Color32 pixel = image->pixel(x, y);

 			// Convert to our desired size, and reconstruct.
-			pixel.r = PixelFormat::convert(pixel.r, 8, rsize);
-			pixel.r = PixelFormat::convert(pixel.r, rsize, 8);
-
-			pixel.g = PixelFormat::convert(pixel.g, 8, gsize);
-			pixel.g = PixelFormat::convert(pixel.g, gsize, 8);
-
-			pixel.b = PixelFormat::convert(pixel.b, 8, bsize);
-			pixel.b = PixelFormat::convert(pixel.b, bsize, 8);
-
-			pixel.a = PixelFormat::convert(pixel.a, 8, asize);
-			pixel.a = PixelFormat::convert(pixel.a, asize, 8);
+			pixel.r = PixelFormat::convert(pixel.r, 8, rsize);
+			pixel.r = PixelFormat::convert(pixel.r, rsize, 8);
+
+			pixel.g = PixelFormat::convert(pixel.g, 8, gsize);
+			pixel.g = PixelFormat::convert(pixel.g, gsize, 8);
+
+			pixel.b = PixelFormat::convert(pixel.b, 8, bsize);
+			pixel.b = PixelFormat::convert(pixel.b, bsize, 8);
+
+			pixel.a = PixelFormat::convert(pixel.a, 8, asize);
+			pixel.a = PixelFormat::convert(pixel.a, asize, 8);

 			// Store color.
 			image->pixel(x, y) = pixel;
@ -155,65 +152,65 @@ void nv::Quantize::Truncate(Image * image, uint rsize, uint gsize, uint bsize, u
 }


-// Error diffusion. Floyd Steinberg.
-void nv::Quantize::FloydSteinberg(Image * image, uint rsize, uint gsize, uint bsize, uint asize)
-{
-	nvCheck(image != NULL);
-	
-	const uint w = image->width();
-	const uint h = image->height();
-	
-	Vector4 * row0 = new Vector4[w+2];
-	Vector4 * row1 = new Vector4[w+2];
-	memset(row0, 0, sizeof(Vector4)*(w+2));
-	memset(row1, 0, sizeof(Vector4)*(w+2));
-	
-	for (uint y = 0; y < h; y++) {
-		for (uint x = 0; x < w; x++) {
-			
-			Color32 pixel = image->pixel(x, y);
-
-			// Add error.
-			pixel.r = clamp(int(pixel.r) + int(row0[1+x].x()), 0, 255);
-			pixel.g = clamp(int(pixel.g) + int(row0[1+x].y()), 0, 255);
-			pixel.b = clamp(int(pixel.b) + int(row0[1+x].z()), 0, 255);
-			pixel.a = clamp(int(pixel.a) + int(row0[1+x].w()), 0, 255);
-			
-			int r = pixel.r;
-			int g = pixel.g;
-			int b = pixel.b;
-			int a = pixel.a;
-
-			// Convert to our desired size, and reconstruct.
-			r = PixelFormat::convert(r, 8, rsize);
-			r = PixelFormat::convert(r, rsize, 8);
-
-			g = PixelFormat::convert(g, 8, gsize);
-			g = PixelFormat::convert(g, gsize, 8);
-
-			b = PixelFormat::convert(b, 8, bsize);
-			b = PixelFormat::convert(b, bsize, 8);
-
-			a = PixelFormat::convert(a, 8, asize);
-			a = PixelFormat::convert(a, asize, 8);
-
-			// Store color.
-			image->pixel(x, y) = Color32(r, g, b, a);
-			
-			// Compute new error.
-			Vector4 diff(float(int(pixel.r) - r), float(int(pixel.g) - g), float(int(pixel.b) - b), float(int(pixel.a) - a));
-			
-			// Propagate new error.
-			row0[1+x+1] += 7.0f / 16.0f * diff;
-			row1[1+x-1] += 3.0f / 16.0f * diff;
-			row1[1+x+0] += 5.0f / 16.0f * diff;
-			row1[1+x+1] += 1.0f / 16.0f * diff;
-		}
-		
-		swap(row0, row1);
-		memset(row1, 0, sizeof(Vector4)*(w+2));
-	}
-	
-	delete [] row0;
-	delete [] row1;
-}
+// Error diffusion. Floyd Steinberg.
+void nv::Quantize::FloydSteinberg(Image * image, uint rsize, uint gsize, uint bsize, uint asize)
+{
+	nvCheck(image != NULL);
+	
+	const uint w = image->width();
+	const uint h = image->height();
+	
+	Vector4 * row0 = new Vector4[w+2];
+	Vector4 * row1 = new Vector4[w+2];
+	memset(row0, 0, sizeof(Vector4)*(w+2));
+	memset(row1, 0, sizeof(Vector4)*(w+2));
+	
+	for (uint y = 0; y < h; y++) {
+		for (uint x = 0; x < w; x++) {
+			
+			Color32 pixel = image->pixel(x, y);
+
+			// Add error.
+			pixel.r = clamp(int(pixel.r) + int(row0[1+x].x()), 0, 255);
+			pixel.g = clamp(int(pixel.g) + int(row0[1+x].y()), 0, 255);
+			pixel.b = clamp(int(pixel.b) + int(row0[1+x].z()), 0, 255);
+			pixel.a = clamp(int(pixel.a) + int(row0[1+x].w()), 0, 255);
+			
+			int r = pixel.r;
+			int g = pixel.g;
+			int b = pixel.b;
+			int a = pixel.a;
+
+			// Convert to our desired size, and reconstruct.
+			r = PixelFormat::convert(r, 8, rsize);
+			r = PixelFormat::convert(r, rsize, 8);
+
+			g = PixelFormat::convert(g, 8, gsize);
+			g = PixelFormat::convert(g, gsize, 8);
+
+			b = PixelFormat::convert(b, 8, bsize);
+			b = PixelFormat::convert(b, bsize, 8);
+
+			a = PixelFormat::convert(a, 8, asize);
+			a = PixelFormat::convert(a, asize, 8);
+
+			// Store color.
+			image->pixel(x, y) = Color32(r, g, b, a);
+			
+			// Compute new error.
+			Vector4 diff(float(int(pixel.r) - r), float(int(pixel.g) - g), float(int(pixel.b) - b), float(int(pixel.a) - a));
+			
+			// Propagate new error.
+			row0[1+x+1] += 7.0f / 16.0f * diff;
+			row1[1+x-1] += 3.0f / 16.0f * diff;
+			row1[1+x+0] += 5.0f / 16.0f * diff;
+			row1[1+x+1] += 1.0f / 16.0f * diff;
+		}
+		
+		swap(row0, row1);
+		memset(row1, 0, sizeof(Vector4)*(w+2));
+	}
+	
+	delete [] row0;
+	delete [] row1;
+}
--- a/src/nvimage/Quantize.h
+++ b/src/nvimage/Quantize.h
@ -3,9 +3,6 @@
 #ifndef NV_IMAGE_QUANTIZE_H
 #define NV_IMAGE_QUANTIZE_H

-#include <nvimage/nvimage.h>
-
-
 namespace nv
 {
 	class Image;
--- a/src/nvmath/Box.h
+++ b/src/nvmath/Box.h
@ -108,7 +108,7 @@ public:
 	float area() const
 	{
 		const Vector3 d = extents();
-		return 8.0f * (d.x()*d.y() + d.x()*d.z() + d.y()*d.z());
+		return 4.0f * (d.x()*d.y() + d.x()*d.z() + d.y()*d.z());
 	}	

 	/// Get the volume of the box.
@ -118,14 +118,6 @@ public:
 		return 8.0f * (d.x() * d.y() * d.z());
 	}
 	
-	/// Return true if the box contains the given point.
-	bool contains(Vector3::Arg p) const
-	{
-		return 
-			m_mins.x() < p.x() && m_mins.y() < p.y() && m_mins.z() < p.z() &&
-			m_maxs.x() > p.x() && m_maxs.y() > p.y() && m_maxs.z() > p.z();
-	}
-
 private:

 	Vector3 m_mins;
@ -133,6 +125,15 @@ private:
 };


+/*
+/// Point inside box test.
+inline bool pointInsideBox(const Box & b, Vector3::Arg p) const
+{
+	return (m_mins.x() < p.x() && m_mins.y() < p.y() && m_mins.z() < p.z() &&
+		m_maxs.x() > p.x() && m_maxs.y() > p.y() && m_maxs.z() > p.z());
+}
+*/
+

 } // nv namespace

--- a/src/nvmath/CMakeLists.txt
+++ b/src/nvmath/CMakeLists.txt
@ -19,7 +19,6 @@ INCLUDE_DIRECTORIES(${CMAKE_CURRENT_SOURCE_DIR})
 ADD_DEFINITIONS(-DNVMATH_EXPORTS)

 IF(NVMATH_SHARED)
-	ADD_DEFINITIONS(-DNVMATH_SHARED=1)
 	ADD_LIBRARY(nvmath SHARED ${MATH_SRCS})
 ELSE(NVMATH_SHARED)
 	ADD_LIBRARY(nvmath ${MATH_SRCS})
--- a/src/nvmath/Matrix.h
+++ b/src/nvmath/Matrix.h
@ -332,7 +332,7 @@ inline Matrix transpose(Matrix::Arg m)
 	Matrix r;
 	for (int i = 0; i < 4; i++)
 	{
-		for (int j = 0; j < 4; j++)
+		for (int j = 0; j < 4; i++)
 		{
 			r(i, j) = m(j, i);
 		}
--- a/src/nvmath/nvmath.h
+++ b/src/nvmath/nvmath.h
@ -48,37 +48,19 @@
 #define IS_NEGATIVE_FLOAT(x)	(IR(x)&SIGN_BITMASK)
 */

-inline double sqrt_assert(const double f)
-{
-	nvDebugCheck(f >= 0.0f);
-	return sqrt(f);
-}
-
-inline float sqrtf_assert(const float f)
+inline float sqrt_assert(const float f)
 {
 	nvDebugCheck(f >= 0.0f);
 	return sqrtf(f);
 }

-inline double acos_assert(const double f)
-{
-	nvDebugCheck(f >= -1.0f && f <= 1.0f);
-	return acos(f);
-}
-
-inline float acosf_assert(const float f)
+inline float acos_assert(const float f)
 {
 	nvDebugCheck(f >= -1.0f && f <= 1.0f);
 	return acosf(f);
 }

-inline double asin_assert(const double f)
-{
-	nvDebugCheck(f >= -1.0f && f <= 1.0f);
-	return asin(f);
-}
-
-inline float asinf_assert(const float f)
+inline float asin_assert(const float f)
 {
 	nvDebugCheck(f >= -1.0f && f <= 1.0f);
 	return asinf(f);
@ -86,11 +68,11 @@ inline float asinf_assert(const float f)

 // Replace default functions with asserting ones.
 #define sqrt sqrt_assert
-#define sqrtf sqrtf_assert
+#define sqrtf sqrt_assert
 #define acos acos_assert
-#define acosf acosf_assert
+#define acosf acos_assert
 #define asin asin_assert
-#define asinf asinf_assert
+#define asinf asin_assert

 #if NV_OS_WIN32
 #include <float.h>
@ -154,11 +136,6 @@ inline float lerp(float f0, float f1, float t)
 	return f0 * s + f1 * t;
 }

-inline float square(float f)
-{
-	return f * f;
-}
-
 } // nv

 #endif // NV_MATH_H
--- a/src/nvtt/CMakeLists.txt
+++ b/src/nvtt/CMakeLists.txt
@ -43,9 +43,8 @@ INCLUDE_DIRECTORIES(${CMAKE_CURRENT_SOURCE_DIR})

 ADD_DEFINITIONS(-DNVTT_EXPORTS)

-IF(NVTT_SHARED)
-	ADD_DEFINITIONS(-DNVTT_SHARED=1)
-	ADD_LIBRARY(nvtt SHARED ${NVTT_SRCS})
+IF(NVTT_SHARED)	
+	ADD_LIBRARY(nvtt SHARED ${DXT_SRCS})
 ELSE(NVTT_SHARED)
 	ADD_LIBRARY(nvtt ${NVTT_SRCS})
 ENDIF(NVTT_SHARED)
--- a/src/nvtt/CompressDXT.cpp
+++ b/src/nvtt/CompressDXT.cpp
@ -205,9 +205,9 @@ void nv::SlowCompressor::compressDXT1(const CompressionOptions::Private & compre
 	ColorBlock rgba;
 	BlockDXT1 block;

-	squish::WeightedClusterFit fit;
+	//squish::WeightedClusterFit fit;
 	//squish::ClusterFit fit;
-	//squish::FastClusterFit fit;
+	squish::FastClusterFit fit;
 	fit.SetMetric(compressionOptions.colorWeight.x(), compressionOptions.colorWeight.y(), compressionOptions.colorWeight.z());

 	for (uint y = 0; y < h; y += 4) {
@ -221,7 +221,7 @@ void nv::SlowCompressor::compressDXT1(const CompressionOptions::Private & compre
 			}
 			else
 			{
-				squish::ColourSet colours((uint8 *)rgba.colors(), 0, true);
+				squish::ColourSet colours((uint8 *)rgba.colors(), 0);
 				fit.SetColourSet(&colours, squish::kDxt1);
 				fit.Compress(&block);
 			}
--- a/src/nvtt/Compressor.cpp
+++ b/src/nvtt/Compressor.cpp
@ -53,7 +53,7 @@ using namespace nvtt;

 namespace
 {
-
+	
 	static int blockSize(Format format)
 	{
 		if (format == Format_DXT1 || format == Format_DXT1a) {
@ -121,13 +121,15 @@ namespace nvtt
 			m_fixedImage = NULL;
 			m_floatImage = image;
 		}
-
+		

 		// Convert linear float image to fixed image ready for compression.
 		void toFixedImage(const InputOptions::Private & inputOptions)
 		{
-			if (m_floatImage != NULL) // apfaffe - We should check that we have a float image, if so convert it!
+			if (this->asFixedImage() == NULL)
 			{
+				nvDebugCheck(m_floatImage != NULL);
+
 				if (inputOptions.isNormalMap || inputOptions.outputGamma == 1.0f)
 				{
 					m_fixedImage = m_floatImage->createImage();
@ -151,7 +153,7 @@ namespace nvtt
 				if (inputOptions.isNormalMap)
 				{
 					// Expand normals to [-1, 1] range.
-					//	floatImage->expandNormals(0);
+				//	floatImage->expandNormals(0);
 				}
 				else if (inputOptions.inputGamma != 1.0f)
 				{
@ -173,12 +175,11 @@ namespace nvtt

 		const Image * asFixedImage() const
 		{
-			// - apfaffe - switched logic to return the 'processed image' rather than the input!
-			if (m_fixedImage != NULL && m_fixedImage.ptr() != NULL)
+			if (m_inputImage != NULL) 
 			{
-				return m_fixedImage.ptr();
+				return m_inputImage;
 			}
-			return m_inputImage;
+			return m_fixedImage.ptr();
 		}

 		Image * asMutableFixedImage()
@ -192,7 +193,7 @@ namespace nvtt
 			return m_fixedImage.ptr();
 		}

-
+		
 	private:
 		const Image * m_inputImage;
 		AutoPtr<Image> m_fixedImage;
@ -206,15 +207,22 @@ Compressor::Compressor() : m(*new Compressor::Private())
 {
 	// CUDA initialization.
 	m.cudaSupported = cuda::isHardwarePresent();
-	m.cudaEnabled = false;
-	m.cudaDevice = -1;
+	m.cudaEnabled = m.cudaSupported;

-	enableCudaAcceleration(m.cudaSupported);
+	if (m.cudaEnabled)
+	{
+		m.cuda = new CudaCompressor();
+
+		if (!m.cuda->isValid())
+		{
+			m.cudaEnabled = false;
+			m.cuda = NULL;
+		}
+	}
 }

 Compressor::~Compressor()
 {
-	enableCudaAcceleration(false);
 	delete &m;
 }

@ -224,33 +232,17 @@ void Compressor::enableCudaAcceleration(bool enable)
 {
 	if (m.cudaSupported)
 	{
-		if (m.cudaEnabled && !enable)
+		m.cudaEnabled = enable;
+	}
+
+	if (m.cudaEnabled && m.cuda == NULL)
+	{
+		m.cuda = new CudaCompressor();
+		
+		if (!m.cuda->isValid())
 		{
 			m.cudaEnabled = false;
 			m.cuda = NULL;
-
-			if (m.cudaDevice != -1)
-			{
-				// Exit device.
-				cuda::exitDevice();
-			}
-		}
-		else if (!m.cudaEnabled && enable)
-		{
-			// Init the CUDA device. This may return -1 if CUDA was already initialized by the app.
-			m.cudaEnabled = cuda::initDevice(&m.cudaDevice);
-
-			if (m.cudaEnabled)
-			{
-				// Create compressor if initialization succeeds.
-				m.cuda = new CudaCompressor();
-
-				// But cleanup if failed.
-				if (!m.cuda->isValid())
-				{
-					enableCudaAcceleration(false);
-				}
-			}
 		}
 	}
 }
@ -291,9 +283,9 @@ bool Compressor::Private::compress(const InputOptions::Private & inputOptions, c
 		if (outputOptions.errorHandler) outputOptions.errorHandler->error(Error_FileOpen);
 		return false;
 	}
-
+	
 	inputOptions.computeTargetExtents();
-
+	
 	// Output DDS header.
 	if (!outputHeader(inputOptions, compressionOptions, outputOptions))
 	{
@ -309,7 +301,7 @@ bool Compressor::Private::compress(const InputOptions::Private & inputOptions, c
 	}

 	outputOptions.closeFile();
-
+	
 	return true;
 }

@ -324,15 +316,15 @@ bool Compressor::Private::outputHeader(const InputOptions::Private & inputOption
 	}

 	DDSHeader header;
-
+	
 	header.setWidth(inputOptions.targetWidth);
 	header.setHeight(inputOptions.targetHeight);
-
+	
 	int mipmapCount = inputOptions.realMipmapCount();
 	nvDebugCheck(mipmapCount > 0);
-
+	
 	header.setMipmapCount(mipmapCount);
-
+	
 	if (inputOptions.textureType == TextureType_2D) {
 		header.setTexture2D();
 	}
@ -340,10 +332,10 @@ bool Compressor::Private::outputHeader(const InputOptions::Private & inputOption
 		header.setTextureCube();
 	}		
 	/*else if (inputOptions.textureType == TextureType_3D) {
-	header.setTexture3D();
-	header.setDepth(inputOptions.targetDepth);
+		header.setTexture3D();
+		header.setDepth(inputOptions.targetDepth);
 	}*/
-
+	
 	if (compressionOptions.format == Format_RGBA)
 	{
 		header.setPitch(computePitch(inputOptions.targetWidth, compressionOptions.bitcount));
@ -352,7 +344,7 @@ bool Compressor::Private::outputHeader(const InputOptions::Private & inputOption
 	else
 	{
 		header.setLinearSize(computeImageSize(inputOptions.targetWidth, inputOptions.targetHeight, inputOptions.targetDepth, compressionOptions.bitcount, compressionOptions.format));
-
+		
 		if (compressionOptions.format == Format_DXT1 || compressionOptions.format == Format_DXT1a) {
 			header.setFourCC('D', 'X', 'T', '1');
 			if (inputOptions.isNormalMap) header.setNormalFlag(true);
@ -375,10 +367,10 @@ bool Compressor::Private::outputHeader(const InputOptions::Private & inputOption
 			if (inputOptions.isNormalMap) header.setNormalFlag(true);
 		}
 	}
-
+	
 	// Swap bytes if necessary.
 	header.swapBytes();
-
+	
 	uint headerSize = 128;
 	if (header.hasDX10Header())
 	{
@ -391,7 +383,7 @@ bool Compressor::Private::outputHeader(const InputOptions::Private & inputOption
 	{
 		outputOptions.errorHandler->error(Error_FileWrite);
 	}
-
+	
 	return writeSucceed;
 }

@ -427,7 +419,7 @@ bool Compressor::Private::compressMipmaps(uint f, const InputOptions::Private &
 				return false;
 			}
 		}
-
+		
 		quantizeMipmap(mipmap, compressionOptions);

 		compressMipmap(mipmap, inputOptions, compressionOptions, outputOptions);
@ -437,7 +429,7 @@ bool Compressor::Private::compressMipmaps(uint f, const InputOptions::Private &
 		h = max(1U, h / 2);
 		d = max(1U, d / 2);
 	}
-
+	
 	return true;
 }

@ -488,7 +480,7 @@ int Compressor::Private::findExactMipmap(const InputOptions::Private & inputOpti
 	{
 		int idx = f * inputOptions.mipmapCount + m;
 		const InputOptions::Private::InputImage & inputImage = inputOptions.images[idx];
-
+		
 		if (inputImage.width == int(w) && inputImage.height == int(h) && inputImage.depth == int(d))
 		{
 			if (inputImage.data != NULL)
@ -543,7 +535,7 @@ void Compressor::Private::downsampleMipmap(Mipmap & mipmap, const InputOptions::
 	mipmap.toFloatImage(inputOptions);

 	const FloatImage * floatImage = mipmap.asFloatImage();
-
+	
 	if (inputOptions.mipmapFilter == MipmapFilter_Box)
 	{
 		// Use fast downsample.
@ -561,7 +553,7 @@ void Compressor::Private::downsampleMipmap(Mipmap & mipmap, const InputOptions::
 		filter.setParameters(inputOptions.kaiserAlpha, inputOptions.kaiserStretch);
 		mipmap.setImage(floatImage->downSample(filter, (FloatImage::WrapMode)inputOptions.wrapMode));
 	}
-
+	
 	// Normalize mipmap.
 	if ((inputOptions.isNormalMap || inputOptions.convertToNormalMap) && inputOptions.normalizeMipmaps)
 	{
@ -589,7 +581,7 @@ void Compressor::Private::processInputImage(Mipmap & mipmap, const InputOptions:
 	if (inputOptions.convertToNormalMap)
 	{
 		mipmap.toFixedImage(inputOptions);
-
+		
 		Vector4 heightScale = inputOptions.heightFactors;
 		mipmap.setImage(createNormalMap(mipmap.asFixedImage(), (FloatImage::WrapMode)inputOptions.wrapMode, heightScale, inputOptions.bumpFrequencyScale));
 	}
@ -697,7 +689,6 @@ bool Compressor::Private::compressMipmap(const Mipmap & mipmap, const InputOptio
 	SlowCompressor slow;
 	slow.setImage(image, inputOptions.alphaMode);

-	const bool useCuda = cudaEnabled && image->width() * image->height() >= 512;

 	if (compressionOptions.format == Format_RGBA || compressionOptions.format == Format_RGB)
 	{
@ -714,29 +705,29 @@ bool Compressor::Private::compressMipmap(const Mipmap & mipmap, const InputOptio
 #endif

 #if defined(HAVE_ATITC)
-			if (compressionOptions.externalCompressor == "ati")
+		if (compressionOptions.externalCompressor == "ati")
+		{
+			atiCompressDXT1(image, outputOptions);
+		}
+		else
+#endif
+		if (compressionOptions.quality == Quality_Fastest)
+		{
+			fast.compressDXT1(outputOptions);
+		}
+		else
+		{
+			if (cudaEnabled)
 			{
-				atiCompressDXT1(image, outputOptions);
+				nvDebugCheck(cudaSupported);
+				cuda->setImage(image, inputOptions.alphaMode);
+				cuda->compressDXT1(compressionOptions, outputOptions);
 			}
 			else
-#endif
-				if (compressionOptions.quality == Quality_Fastest)
-				{
-					fast.compressDXT1(outputOptions);
-				}
-				else
-				{
-					if (useCuda)
-					{
-						nvDebugCheck(cudaSupported);
-						cuda->setImage(image, inputOptions.alphaMode);
-						cuda->compressDXT1(compressionOptions, outputOptions);
-					}
-					else
-					{
-						slow.compressDXT1(compressionOptions, outputOptions);
-					}
-				}
+			{
+				slow.compressDXT1(compressionOptions, outputOptions);
+			}
+		}
 	}
 	else if (compressionOptions.format == Format_DXT1a)
 	{
@ -746,7 +737,7 @@ bool Compressor::Private::compressMipmap(const Mipmap & mipmap, const InputOptio
 		}
 		else
 		{
-			if (useCuda)
+			if (cudaEnabled)
 			{
 				nvDebugCheck(cudaSupported);
 				/*cuda*/slow.compressDXT1a(compressionOptions, outputOptions);
@ -765,7 +756,7 @@ bool Compressor::Private::compressMipmap(const Mipmap & mipmap, const InputOptio
 		}
 		else
 		{
-			if (useCuda)
+			if (cudaEnabled)
 			{
 				nvDebugCheck(cudaSupported);
 				cuda->setImage(image, inputOptions.alphaMode);
@ -785,7 +776,7 @@ bool Compressor::Private::compressMipmap(const Mipmap & mipmap, const InputOptio
 		}
 		else
 		{
-			if (useCuda)
+			if (cudaEnabled)
 			{
 				nvDebugCheck(cudaSupported);
 				cuda->setImage(image, inputOptions.alphaMode);
@ -827,27 +818,27 @@ int Compressor::Private::estimateSize(const InputOptions::Private & inputOptions
 	const uint bitCount = compressionOptions.bitcount;

 	inputOptions.computeTargetExtents();
-
+	
 	uint mipmapCount = inputOptions.realMipmapCount();
-
+	
 	int size = 0;
-
+	
 	for (uint f = 0; f < inputOptions.faceCount; f++)
 	{
 		uint w = inputOptions.targetWidth;
 		uint h = inputOptions.targetHeight;
 		uint d = inputOptions.targetDepth;
-
+		
 		for (uint m = 0; m < mipmapCount; m++)
 		{
 			size += computeImageSize(w, h, d, bitCount, format);
-
+			
 			// Compute extents of next mipmap:
 			w = max(1U, w / 2);
 			h = max(1U, h / 2);
 			d = max(1U, d / 2);
 		}
 	}
-
+	
 	return size;
 }
--- a/src/nvtt/Compressor.h
+++ b/src/nvtt/Compressor.h
@ -63,12 +63,10 @@ namespace nvtt
 		bool compressMipmap(const Mipmap & mipmap, const InputOptions::Private & inputOptions, const CompressionOptions::Private & compressionOptions, const OutputOptions::Private & outputOptions) const;


-
 	public:

 		bool cudaSupported;
 		bool cudaEnabled;
-		int cudaDevice;

 		nv::AutoPtr<nv::CudaCompressor> cuda;

--- a/src/nvtt/InputOptions.cpp
+++ b/src/nvtt/InputOptions.cpp
@ -94,7 +94,7 @@ void InputOptions::reset()
 	m.textureType = TextureType_2D;
 	m.inputFormat = InputFormat_BGRA_8UB;

-	m.alphaMode = AlphaMode_None;
+	m.alphaMode = AlphaMode_Transparency;

 	m.inputGamma = 2.2f;
 	m.outputGamma = 2.2f;
--- a/src/nvtt/SingleColorLookup.h
+++ b/src/nvtt/SingleColorLookup.h
--- a/src/nvtt/cuda/CompressKernel.cu
+++ b/src/nvtt/cuda/CompressKernel.cu
@ -191,9 +191,6 @@ __device__ void loadColorBlock(const uint * image, float3 colors[16], float3 sum

 		*sameColor = (axis == make_float3(0, 0, 0));

-		// Single color compressor needs unweighted colors.
-		if (*sameColor) colors[idx] = rawColors[idx];
-
 		dps[idx] = dot(rawColors[idx], axis);
 		
 #if __DEVICE_EMULATION__
--- a/src/nvtt/cuda/CudaCompressDXT.cpp
+++ b/src/nvtt/cuda/CudaCompressDXT.cpp
@ -37,7 +37,7 @@


 #if defined HAVE_CUDA
-#include <cuda_runtime_api.h>
+#include <cuda_runtime.h>
 #endif

 #include <time.h>
--- a/src/nvtt/cuda/CudaMath.h
+++ b/src/nvtt/cuda/CudaMath.h
@ -127,21 +127,8 @@ inline __device__ __host__ float3 normalize(float3 v)
 inline __device__ __host__ float3 firstEigenVector( float matrix[6] )
 {
 	// 8 iterations seems to be more than enough.
-
-	float3 row0 = make_float3(matrix[0], matrix[1], matrix[2]);
-	float3 row1 = make_float3(matrix[1], matrix[3], matrix[4]);
-	float3 row2 = make_float3(matrix[2], matrix[4], matrix[5]);

-	float r0 = dot(row0, row0);
-	float r1 = dot(row1, row1);
-	float r2 = dot(row2, row2);
-
-	float3 v;
-	if (r0 > r1 && r0 > r2) v = row0;
-	else if (r1 > r2) v = row1;
-	else v = row2;
-
-	//float3 v = make_float3(1.0f, 1.0f, 1.0f);
+	float3 v = make_float3(1.0f, 1.0f, 1.0f);
 	for(int i = 0; i < 8; i++) {
 		float x = v.x * matrix[0] + v.y * matrix[1] + v.z * matrix[2];
 		float y = v.x * matrix[1] + v.y * matrix[3] + v.z * matrix[4];
@ -161,7 +148,7 @@ inline __device__ bool singleColor(const float3 * colors)
 	bool sameColor = false;
 	for (int i = 0; i < 16; i++)
 	{
-		sameColor &= (colors[i] == colors[0]);
+		sameColor &= (colors[idx] == colors[0]);
 	}
 	return sameColor;
 #else
--- a/src/nvtt/cuda/CudaUtils.cpp
+++ b/src/nvtt/cuda/CudaUtils.cpp
@ -1,300 +1,128 @@
-// Copyright NVIDIA Corporation 2007 -- Ignacio Castano <icastano@nvidia.com>
-// 
-// Permission is hereby granted, free of charge, to any person
-// obtaining a copy of this software and associated documentation
-// files (the "Software"), to deal in the Software without
-// restriction, including without limitation the rights to use,
-// copy, modify, merge, publish, distribute, sublicense, and/or sell
-// copies of the Software, and to permit persons to whom the
-// Software is furnished to do so, subject to the following
-// conditions:
-// 
-// The above copyright notice and this permission notice shall be
-// included in all copies or substantial portions of the Software.
-// 
-// THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND,
-// EXPRESS OR IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES
-// OF MERCHANTABILITY, FITNESS FOR A PARTICULAR PURPOSE AND
-// NONINFRINGEMENT. IN NO EVENT SHALL THE AUTHORS OR COPYRIGHT
-// HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER LIABILITY,
-// WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING
-// FROM, OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR
-// OTHER DEALINGS IN THE SOFTWARE.
-
-#include <nvcore/Debug.h>
-#include <nvcore/Library.h>
-#include "CudaUtils.h"
-
-#if defined HAVE_CUDA
-#include <cuda.h>
-#include <cuda_runtime_api.h>
-#endif
-
-using namespace nv;
-using namespace cuda;
-
-/* @@ Move this to win32 utils or somewhere else.
-#if NV_OS_WIN32
-
-#define WINDOWS_LEAN_AND_MEAN
-#include <windows.h>
-
-static bool isWindowsVista()
-{
-OSVERSIONINFO osvi;
-osvi.dwOSVersionInfoSize = sizeof(OSVERSIONINFO);
-
-::GetVersionEx(&osvi);
-return osvi.dwMajorVersion >= 6;
-}
-
-
-typedef BOOL (WINAPI *LPFN_ISWOW64PROCESS) (HANDLE, PBOOL);
-
-static bool isWow32()
-{
-LPFN_ISWOW64PROCESS fnIsWow64Process = (LPFN_ISWOW64PROCESS)GetProcAddress(GetModuleHandle("kernel32"), "IsWow64Process");
-
-BOOL bIsWow64 = FALSE;
-
-if (NULL != fnIsWow64Process)
-{
-if (!fnIsWow64Process(GetCurrentProcess(), &bIsWow64))
-{
-// Assume 32 bits.
-return true;
-}
-}
-
-return !bIsWow64;
-}
-
-#endif
-*/
-
-
-static bool isCudaDriverAvailable(int version)
-{
-#if defined HAVE_CUDA
-#if NV_OS_WIN32
-	Library nvcuda("nvcuda.dll");
-#else
-	Library nvcuda(NV_LIBRARY_NAME(cuda));
-#endif
-
-	if (!nvcuda.isValid())
-	{
-		nvDebug("*** CUDA driver not found.\n");
-		return false;
-	}
-
-	if (version >= 2000)
-	{
-		void * address = nvcuda.bindSymbol("cuStreamCreate");
-		if (address == NULL) {
-			nvDebug("*** CUDA driver version < 2.0.\n");
-			return false;
-		}
-	}
-
-	if (version >= 2010)
-	{
-		void * address = nvcuda.bindSymbol("cuModuleLoadDataEx");
-		if (address == NULL) {
-			nvDebug("*** CUDA driver version < 2.1.\n");
-			return false;
-		}
-	}
-
-	if (version >= 2020)
-	{
-		typedef CUresult (CUDAAPI * PFCU_DRIVERGETVERSION)(int * version);
-
-		PFCU_DRIVERGETVERSION driverGetVersion = (PFCU_DRIVERGETVERSION)nvcuda.bindSymbol("cuDriverGetVersion");
-		if (driverGetVersion == NULL) {
-			nvDebug("*** CUDA driver version < 2.2.\n");
-			return false;
-		}
-
-		int driverVersion;
-		CUresult err = driverGetVersion(&driverVersion);
-		if (err != CUDA_SUCCESS) {
-			nvDebug("*** Error querying driver version: '%s'.\n", cudaGetErrorString((cudaError_t)err));
-			return false;
-		}
-
-		return driverVersion >= version;
-	}
-#endif // HAVE_CUDA
-
-	return true;
-}
-
-
-/// Determine if CUDA is available.
-bool nv::cuda::isHardwarePresent()
-{
-#if defined HAVE_CUDA
-	// Make sure that CUDA driver matches CUDA runtime.
-	if (!isCudaDriverAvailable(CUDART_VERSION))
-	{
-		nvDebug("CUDA driver not available for CUDA runtime %d\n", CUDART_VERSION);
-		return false;
-	}
-
-	int count = deviceCount();
-	if (count == 1)
-	{
-		// Make sure it's not an emulation device.
-		cudaDeviceProp deviceProp;
-		cudaGetDeviceProperties(&deviceProp, 0);
-
-		// deviceProp.name != Device Emulation (CPU)
-		if (deviceProp.major == -1 || deviceProp.minor == -1)
-		{
-			return false;
-		}
-	}
-
-	// @@ Make sure that warp size == 32
-
-	// @@ Make sure available GPU is faster than the CPU.
-
-	return count > 0;
-#else
-	return false;
-#endif
-}
-
-/// Get number of CUDA enabled devices.
-int nv::cuda::deviceCount()
-{
-#if defined HAVE_CUDA
-	int gpuCount = 0;
-
-	cudaError_t result = cudaGetDeviceCount(&gpuCount);
-
-	if (result == cudaSuccess)
-	{
-		return gpuCount;
-	}
-#endif
-	return 0;
-}
-
-
-// Make sure device meets requirements:
-// - Not an emulation device.
-// - Not an integrated device?
-// - Faster than CPU.
-bool nv::cuda::isValidDevice(int i)
-{
-#if defined HAVE_CUDA
-
-	cudaDeviceProp device_properties;
-	cudaGetDeviceProperties(&device_properties, i);
-	int gflops = device_properties.multiProcessorCount * device_properties.clockRate;
-
-	if (device_properties.major == -1 || device_properties.minor == -1) {
-		// Emulation device.
-		return false;
-	}
-
-#if CUDART_VERSION >= 2030 // 2.3
-	/*if (device_properties.integrated)
-	{
-		// Integrated devices.
-		return false;
-	}*/
-#endif
-
-	return true;
-#else
-	return false;
-#endif
-}
-
-int nv::cuda::getFastestDevice()
-{
-	int max_gflops_device = -1;
-#if defined HAVE_CUDA
-	int max_gflops = 0;
-
-	const int device_count = deviceCount();
-	for (int i = 0; i < device_count; i++)
-	{
-		if (isValidDevice(i))
-		{
-			cudaDeviceProp device_properties;
-			cudaGetDeviceProperties(&device_properties, i);
-			int gflops = device_properties.multiProcessorCount * device_properties.clockRate;
-
-			if (gflops > max_gflops)
-			{
-				max_gflops = gflops;
-				max_gflops_device = i;
-			}
-		}
-	}
-#endif
-	return max_gflops_device;
-}
-
-
-/// Activate the given devices.
-bool nv::cuda::initDevice(int * device_ptr)
-{
-	nvDebugCheck(device_ptr != NULL);
-#if defined HAVE_CUDA
-
-#if CUDART_VERSION >= 2030 // 2.3
-
-	// Set device flags to yield in order to play nice with other threads and to find out if CUDA was already active.
-	cudaError_t resul = cudaSetDeviceFlags(cudaDeviceScheduleYield);
-
-#endif
-
-	int device = getFastestDevice();
-
-	if (device == -1)
-	{
-		// No device is fast enough.
-		*device_ptr = -1;
-		return false;
-	}
-
-	// Select CUDA device.
-	cudaError_t result = cudaSetDevice(device);
-
-	if (result == cudaErrorSetOnActiveProcess)
-	{
-		int device;
-		result = cudaGetDevice(&device);
-
-		*device_ptr = -1;  // No device to cleanup.
-		return isValidDevice(device); // Return true if device is valid.
-	}
-	else if (result != cudaSuccess)
-	{
-		nvDebug("*** CUDA Error: %s\n", cudaGetErrorString(result));
-		*device_ptr = -1;
-		return false;
-	}
-
-	*device_ptr = device;
-	return true;
-#else
-	return false;
-#endif
-}
-
-void nv::cuda::exitDevice()
-{
-#if defined HAVE_CUDA
-	cudaError_t result = cudaThreadExit();
-
-	if (result != cudaSuccess) {
-		nvDebug("*** CUDA Error: %s\n", cudaGetErrorString(result));
-	}
-#endif
-}
+// Copyright NVIDIA Corporation 2007 -- Ignacio Castano <icastano@nvidia.com>
+// 
+// Permission is hereby granted, free of charge, to any person
+// obtaining a copy of this software and associated documentation
+// files (the "Software"), to deal in the Software without
+// restriction, including without limitation the rights to use,
+// copy, modify, merge, publish, distribute, sublicense, and/or sell
+// copies of the Software, and to permit persons to whom the
+// Software is furnished to do so, subject to the following
+// conditions:
+// 
+// The above copyright notice and this permission notice shall be
+// included in all copies or substantial portions of the Software.
+// 
+// THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND,
+// EXPRESS OR IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES
+// OF MERCHANTABILITY, FITNESS FOR A PARTICULAR PURPOSE AND
+// NONINFRINGEMENT. IN NO EVENT SHALL THE AUTHORS OR COPYRIGHT
+// HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER LIABILITY,
+// WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING
+// FROM, OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR
+// OTHER DEALINGS IN THE SOFTWARE.
+
+#include <nvcore/Debug.h>
+#include "CudaUtils.h"
+
+#if defined HAVE_CUDA
+#include <cuda_runtime.h>
+#endif
+
+using namespace nv;
+using namespace cuda;
+
+#if NV_OS_WIN32
+
+#define WINDOWS_LEAN_AND_MEAN
+#include <windows.h>
+
+static bool isWindowsVista()
+{
+	OSVERSIONINFO osvi;
+	osvi.dwOSVersionInfoSize = sizeof(OSVERSIONINFO);
+
+	::GetVersionEx(&osvi);
+	return osvi.dwMajorVersion >= 6;
+}
+
+
+typedef BOOL (WINAPI *LPFN_ISWOW64PROCESS) (HANDLE, PBOOL);
+
+static bool isWow32()
+{
+	LPFN_ISWOW64PROCESS fnIsWow64Process = (LPFN_ISWOW64PROCESS)GetProcAddress(GetModuleHandle("kernel32"), "IsWow64Process");
+
+    BOOL bIsWow64 = FALSE;
+ 
+    if (NULL != fnIsWow64Process)
+    {
+        if (!fnIsWow64Process(GetCurrentProcess(), &bIsWow64))
+        {
+			// Assume 32 bits.
+            return true;
+        }
+    }
+
+    return !bIsWow64;
+}
+
+#endif
+
+
+/// Determine if CUDA is available.
+bool nv::cuda::isHardwarePresent()
+{
+#if defined HAVE_CUDA
+#if NV_OS_WIN32
+	//if (isWindowsVista()) return false;
+	//if (isWindowsVista() || !isWow32()) return false;
+#endif
+	int count = deviceCount();
+	if (count == 1)
+	{
+		// Make sure it's not an emulation device.
+		cudaDeviceProp deviceProp;
+		cudaGetDeviceProperties(&deviceProp, 0);
+
+		// deviceProp.name != Device Emulation (CPU)
+		if (deviceProp.major == -1 || deviceProp.minor == -1)
+		{
+			return false;
+		}
+
+		// @@ Make sure that warp size == 32
+	}
+
+	return count > 0;
+#else
+	return false;
+#endif
+}
+
+/// Get number of CUDA enabled devices.
+int nv::cuda::deviceCount()
+{
+#if defined HAVE_CUDA
+	int gpuCount = 0;
+
+	cudaError_t result = cudaGetDeviceCount(&gpuCount);
+
+	if (result == cudaSuccess)
+	{
+		return gpuCount;
+	}
+#endif
+	return 0;
+}
+
+/// Activate the given devices.
+bool nv::cuda::setDevice(int i)
+{
+	nvCheck(i < deviceCount());
+#if defined HAVE_CUDA
+	cudaError_t result = cudaSetDevice(i);
+	return result == cudaSuccess;
+#else
+	return false;
+#endif
+}
--- a/src/nvtt/cuda/CudaUtils.h
+++ b/src/nvtt/cuda/CudaUtils.h
@ -31,11 +31,7 @@ namespace nv
 	{
 		bool isHardwarePresent();
 		int deviceCount();
-		int getFastestDevice();
-		bool isValidDevice(int i);
-
-		bool initDevice(int * device_ptr);
-		void exitDevice();
+		bool setDevice(int i);
 	};
 	
 } // nv namespace
--- a/src/nvtt/nvtt.h
+++ b/src/nvtt/nvtt.h
@ -73,7 +73,7 @@ namespace nvtt
 		Format_DXT1a,   // DXT1 with binary alpha.
 		Format_DXT3,
 		Format_DXT5,
-		Format_DXT5n,   // Compressed HILO: R=1, G=y, B=0, A=x
+		Format_DXT5n,   // Compressed HILO: R=0, G=x, B=0, A=y
 		
 		// DX10 formats.
 		Format_BC1 = Format_DXT1,
@ -194,7 +194,7 @@ namespace nvtt
 		// Describe the format of the input.
 		NVTT_API void setFormat(InputFormat format);
 		
-		// Set the way the input alpha channel is interpreted.
+		// Set the way the input alpha channel is interpreted. @@ Not implemented!
 		NVTT_API void setAlphaMode(AlphaMode alphaMode);
 		
 		// Set gamma settings.
--- a/src/nvtt/squish/CMakeLists.txt
+++ b/src/nvtt/squish/CMakeLists.txt
@ -1,8 +1,13 @@
 PROJECT(squish)
+ENABLE_TESTING()

 INCLUDE_DIRECTORIES(${CMAKE_CURRENT_SOURCE_DIR})

 SET(SQUISH_SRCS
+#	alpha.cpp
+#	alpha.h
+#	clusterfit.cpp
+#	clusterfit.h
 	fastclusterfit.cpp
 	fastclusterfit.h
 	weightedclusterfit.cpp
@ -16,13 +21,32 @@ SET(SQUISH_SRCS
 	config.h
 	maths.cpp
 	maths.h
+#	rangefit.cpp
+#	rangefit.h
+#	singlecolourfit.cpp
+#	singlecolourfit.h
+#	singlecolourlookup.inl
+#	squish.cpp
+#	squish.h
 	simd.h
 	simd_sse.h
 	simd_ve.h)

 ADD_LIBRARY(squish STATIC ${SQUISH_SRCS})

-IF(CMAKE_COMPILER_IS_GNUCXX)
-	SET_TARGET_PROPERTIES(squish PROPERTIES COMPILE_FLAGS -fPIC)
-ENDIF(CMAKE_COMPILER_IS_GNUCXX)
+# libpng
+#FIND_PACKAGE(PNG)
+
+#IF(PNG_FOUND)
+#	INCLUDE_DIRECTORIES(${PNG_INCLUDE_DIR})
+#	ADD_EXECUTABLE(squishpng extra/squishpng.cpp)
+#	TARGET_LINK_LIBRARIES(squishpng squish ${PNG_LIBRARY})
+#ENDIF(PNG_FOUND)
+
+##ADD_EXECUTABLE(squishgen extra/squishgen.cpp)
+
+#ADD_EXECUTABLE(squishtest extra/squishtest.cpp)
+#TARGET_LINK_LIBRARIES(squishtest squish)
+
+#ADD_TEST(SQUISHTEST squishtest)

--- a/src/nvtt/squish/maths.cpp
+++ b/src/nvtt/squish/maths.cpp
@ -24,7 +24,6 @@
   -------------------------------------------------------------------------- */
   
 #include "maths.h"
-#include "simd.h"
 #include <cfloat>

 namespace squish {
@ -61,61 +60,12 @@ Sym3x3 ComputeWeightedCovariance( int n, Vec3 const* points, float const* weight
 }


-static Vec3 EstimatePrincipleComponent( Sym3x3 const& matrix )
-{
-	Vec3 const row0(matrix[0], matrix[1], matrix[2]);
-	Vec3 const row1(matrix[1], matrix[3], matrix[4]);
-	Vec3 const row2(matrix[2], matrix[4], matrix[5]);
-
-	float r0 = Dot(row0, row0);
-	float r1 = Dot(row1, row1);
-	float r2 = Dot(row2, row2);
-
-	if (r0 > r1 && r0 > r2) return row0;
-	if (r1 > r2) return row1;
-	return row2;
-}
-
-
-#define POWER_ITERATION_COUNT   8
-
-#if SQUISH_USE_SIMD
-
 Vec3 ComputePrincipleComponent( Sym3x3 const& matrix )
 {
-	Vec4 const row0( matrix[0], matrix[1], matrix[2], 0.0f );
-	Vec4 const row1( matrix[1], matrix[3], matrix[4], 0.0f );
-	Vec4 const row2( matrix[2], matrix[4], matrix[5], 0.0f );
+	const int NUM = 8;

-	//Vec4 v = VEC4_CONST( 1.0f );
-	//Vec4 v = row0; // row1, row2
-
-	Vec3 v3 = EstimatePrincipleComponent( matrix );
-	Vec4 v( v3.X(), v3.Y(), v3.Z(), 0.0f );
-
-	for( int i = 0; i < POWER_ITERATION_COUNT; ++i )
-	{
-		// matrix multiply
-		Vec4 w = row0*v.SplatX();
-		w = MultiplyAdd(row1, v.SplatY(), w);
-		w = MultiplyAdd(row2, v.SplatZ(), w);
-
-		// get max component from xyz in all channels
-		Vec4 a = Max(w.SplatX(), Max(w.SplatY(), w.SplatZ()));
-
-		// divide through and advance
-		v = w*Reciprocal(a);
-	}
-	return v.GetVec3();
-}
-
-#else
-
-Vec3 ComputePrincipleComponent( Sym3x3 const& matrix )
-{
-	Vec3 v = EstimatePrincipleComponent( matrix );
-	for (int i = 0; i < POWER_ITERATION_COUNT; i++)
-	{
+	Vec3 v(1, 1, 1);
+	for(int i = 0; i < NUM; i++) {
 		float x = v.X() * matrix[0] + v.Y() * matrix[1] + v.Z() * matrix[2];
 		float y = v.X() * matrix[1] + v.Y() * matrix[3] + v.Z() * matrix[4];
 		float z = v.X() * matrix[2] + v.Y() * matrix[4] + v.Z() * matrix[5];
@ -132,6 +82,5 @@ Vec3 ComputePrincipleComponent( Sym3x3 const& matrix )
 	return v;
 }

-#endif

 } // namespace squish
--- a/src/nvtt/squish/weightedclusterfit.cpp
+++ b/src/nvtt/squish/weightedclusterfit.cpp
--- a/src/nvtt/tools/compress.cpp
+++ b/src/nvtt/tools/compress.cpp
@ -87,10 +87,7 @@ struct MyErrorHandler : public nvtt::ErrorHandler
 {
 	virtual void error(nvtt::Error e)
 	{
-#if _DEBUG
 		nvDebugBreak();
-#endif
-		printf("Error: '%s'\n", nvtt::errorString(e));
 	}
 };

@ -134,7 +131,6 @@ int main(int argc, char *argv[])
 	MyAssertHandler assertHandler;
 	MyMessageHandler messageHandler;

-	bool alpha = false;
 	bool normal = false;
 	bool color2normal = false;
 	bool wrapRepeat = false;
@ -158,10 +154,6 @@ int main(int argc, char *argv[])
 		if (strcmp("-color", argv[i]) == 0)
 		{
 		}
-		else if (strcmp("-alpha", argv[i]) == 0)
-		{
-			alpha = true;
-		}
 		else if (strcmp("-normal", argv[i]) == 0)
 		{
 			normal = true;
@ -262,12 +254,7 @@ int main(int argc, char *argv[])
 		}
 	}

-	const uint version = nvtt::version();
-	const uint major = version / 100;
-	const uint minor = version % 100;
-	
-
-	printf("NVIDIA Texture Tools %u.%u - Copyright NVIDIA Corporation 2007\n\n", major, minor);
+	printf("NVIDIA Texture Tools - Copyright NVIDIA Corporation 2007\n\n");

 	if (input.isNull())
 	{
@ -275,7 +262,6 @@ int main(int argc, char *argv[])
 		
 		printf("Input options:\n");
 		printf("  -color   \tThe input image is a color map (default).\n");
-		printf("  -alpha     \tThe input image has an alpha channel used for transparency.\n");		
 		printf("  -normal  \tThe input image is a normal map.\n");
 		printf("  -tonormal\tConvert input to normal map.\n");
 		printf("  -clamp   \tClamp wrapping mode (default).\n");
@ -295,7 +281,7 @@ int main(int argc, char *argv[])
 		printf("  -bc4     \tBC4 format (ATI1)\n");
 		printf("  -bc5     \tBC5 format (3Dc/ATI2)\n\n");
 		
-		return EXIT_FAILURE;
+		return 1;
 	}

 	// @@ Make sure input file exists.
@ -310,13 +296,13 @@ int main(int argc, char *argv[])
 		if (!dds.isValid())
 		{
 			fprintf(stderr, "The file '%s' is not a valid DDS file.\n", input.str());
-			return EXIT_FAILURE;
+			return 1;
 		}
 		
 		if (!dds.isSupported() || dds.isTexture3D())
 		{
 			fprintf(stderr, "The file '%s' is not a supported DDS file.\n", input.str());
-			return EXIT_FAILURE;
+			return 1;
 		}
 		
 		uint faceCount;
@ -353,7 +339,7 @@ int main(int argc, char *argv[])
 		if (!image.load(input))
 		{
 			fprintf(stderr, "The file '%s' is not a supported image type.\n", input.str());
-			return EXIT_FAILURE;
+			return 1;
 		}
 		
 		inputOptions.setTextureLayout(nvtt::TextureType_2D, image.width(), image.height());
@ -369,15 +355,6 @@ int main(int argc, char *argv[])
 		inputOptions.setWrapMode(nvtt::WrapMode_Clamp);
 	}

-	if (alpha)
-	{
-		inputOptions.setAlphaMode(nvtt::AlphaMode_Transparency);
-	}
-	else
-	{
-		inputOptions.setAlphaMode(nvtt::AlphaMode_None);
-	}
-
 	if (normal)
 	{
 		setNormalMap(inputOptions);
@ -419,28 +396,33 @@ int main(int argc, char *argv[])
 		compressionOptions.setExternalCompressor(externalCompressor);
 	}

+	if (format == nvtt::Format_RGB)
+	{
+		compressionOptions.setQuantization(true, false, false);
+		//compressionOptions.setPixelFormat(16, 0xF000, 0x0F00, 0x00F0, 0x000F);
+		compressionOptions.setPixelFormat(16, 
+			0x0F00, 
+			0x00F0, 
+			0x000F, 
+			0xF000);
+		//	0x003F0000, 
+		//	0x00003F00, 
+		//	0x0000003F, 
+		//	0x3F000000);
+	}
+
 	
 	MyErrorHandler errorHandler;
 	MyOutputHandler outputHandler(output);
 	if (outputHandler.stream->isError())
 	{
 		fprintf(stderr, "Error opening '%s' for writting\n", output.str());
-		return EXIT_FAILURE;
+		return 1;
 	}

 	nvtt::Compressor compressor;
 	compressor.enableCudaAcceleration(!nocuda);

-	printf("CUDA acceleration ");
-	if (compressor.isCudaAccelerationEnabled())
-	{
-		printf("ENABLED\n\n");
-	}
-	else
-	{
-		printf("DISABLED\n\n");
-	}
-	
 	outputHandler.setTotal(compressor.estimateSize(inputOptions, compressionOptions));
 	outputHandler.setDisplayProgress(!silent);

@ -453,16 +435,27 @@ int main(int argc, char *argv[])
 //	fflush(stdout);
 //	getchar();

+/*	LARGE_INTEGER temp;
+	QueryPerformanceFrequency((LARGE_INTEGER*) &temp);
+	double freq = ((double) temp.QuadPart) / 1000.0;
+
+    LARGE_INTEGER start_time;
+    QueryPerformanceCounter((LARGE_INTEGER*) &start_time);
+*/
 	clock_t start = clock();
 	
-	if (!compressor.process(inputOptions, compressionOptions, outputOptions))
-	{
-		return EXIT_FAILURE;
-	}
+	compressor.process(inputOptions, compressionOptions, outputOptions);
+/*
+	LARGE_INTEGER end_time;
+	QueryPerformanceCounter((LARGE_INTEGER*) &end_time);
+
+	float diff_time = (float) (((double) end_time.QuadPart - (double) start_time.QuadPart) / freq);
+	printf("\rtime taken: %.3f seconds\n", diff_time/1000);
+*/

 	clock_t end = clock();
 	printf("\rtime taken: %.3f seconds\n", float(end-start) / CLOCKS_PER_SEC);
 	
-	return EXIT_SUCCESS;
+	return 0;
 }

--- a/src/nvtt/tools/imgdiff.cpp
+++ b/src/nvtt/tools/imgdiff.cpp
@ -84,7 +84,7 @@ struct Error
 	{
 		mabse /= samples;
 		mse /= samples;
-		rmse = sqrtf(mse);
+		rmse = sqrt(mse);
 		psnr = (rmse == 0) ? 999.0f : 20.0f * log10(255.0f / rmse);
 	}

@ -134,7 +134,7 @@ struct NormalError
 		{
 			ade /= samples;
 			mse /= samples * 3;
-			rmse = sqrtf(mse);
+			rmse = sqrt(mse);
 			psnr = (rmse == 0) ? 999.0f : 20.0f * log10(255.0f / rmse);
 		}
 	}
--- a/src/nvtt/tools/resize.cpp
+++ b/src/nvtt/tools/resize.cpp
@ -113,7 +113,7 @@ int main(int argc, char *argv[])
 				((nv::KaiserFilter *)filter.ptr())->setParameters(4.0f, 1.0f);
 			}
 		}
-		else if (strcmp("-w", argv[i]) == 0)
+		else if (strcmp("-f", argv[i]) == 0)
 		{
 			if (i+1 == argc) break;
 			i++;
 @ -1 +1 @@
 .0.8
 .0.4