If a shape is considered to be a simple closed smooth curve in the complex plane, meaning that they have derivatives of all orders and that do not cross with themselves, then the Riemannian mapping theorem states that it is possible to map the unit disc conformally (meaning that infinitesimal angle between each two crossing curves is equal to the infinitesimal angle between the transformed curves) to the interior of any such shape. The conformal transformation is unique up to any preceding Möbius transformations mapping the unit disc to itself, this is, maps of the form
$$z \rightarrow \frac{az + b}{\bar{b}z + \bar{a}}$$
If we identify $\nabla_{-}$ with the interior of the shape, and $\Gamma_{-}$ with the interior of the unit disk, then by the Riemannian mapping theorem there exists a map
$$\Phi_{-} : \nabla_{-} \rightarrow \Gamma_{-}$$
With the previous characteristics. Furthermore, if we identify $\nabla_{+}$ with the exterior of the shape, and $\Gamma_{+}$ with the exterior of the unit disk, the same conditions apply
$$\Phi_{+} : \nabla_{+} \rightarrow \Gamma_{+}$$
Now, we construct
$$\Psi = \Phi_{+} \circ \Phi_{-}$$
This provides a fingerprint of the shape unique up to a Moebius transformation.
In practice, we use the Swarz-Christoffel Matlab toolbox, having a complex variable with the given vertices coordinates in $pol$, we compute the map $m$ by
pol=polygon(dots);
fc=diskmap(pol);
ex=extermap(pol);
iex=scmapinv(ex);
m=composite(fc,iex);
To extract the fingerprint, we need to evaluate the map in the $z$ coordinate given the angle, so that $z=exp(-i \theta)$, and the evaluation is $m(z)$ for all $\theta \in [0,2\pi)$. Now, the fingerprint is the angle of a given evaluation $m(z)$, this is, $\hat{\theta}=-\frac{log(m(z))}{i}$. plotting $(\theta,\hat{\theta})$, we see the fingerprint.
Checkout the paper here.
Thursday, May 10, 2012
Tuesday, May 1, 2012
Denoising by convolution using CuFFT
In this post we will create a CUDA program that removes some noise with simple isotropic diffusion. We are going to diffuse the function (image) according to the heat equation
$$u_t = \Delta u$$
Where $\Delta$ is the laplacian operator and $u_t$ is the derivative with respect time. The solution at time $t$ is
$$u(x) = \int G_t (x-y) u_o(y) dy$$
Where $x,y\in \mathbb{R}^2$, $u_0$ is the initial noisy image and $G_t$ is the bivariate (normalized) gaussian kernel with standard deviation $t$.
Since a convolution is equivalent to a multiplication of the Fourier transforms as below (by the convolution theorem, and transforming back), we can use the CuFFT library to perform these steps.
$$\hat{u}(\omega) = \hat{G_t}(\omega)\hat{u_o}(\omega)$$
With this project, I want to hunt down two birds with one shot. I want to make a managed code application with Windows forms that reads and visualizes a JPG file, adds noise and then blurs it to remove some noise.
Now, you must know that you cannot mix managed code and CUDA. The solution to this is to separate the CUDA computation and the presentation application in two different parts. We will encapsulate the CUDA computations into a DLL and the Windows program in a Windows Forms application. Both will be part of the same Visual Studio solution so everything will work in the final step. The Windows Forms will be the main project within the VS solution.
Under the DLL project, you must set up CUDA libraries. For the compiler to look for CUDA and CUDAutils header files, we must add
Then our DLL function takes a matrix of values and performs the Fourier transform of both (previously padded) the input data and the kernel. We can precompute the kernel with Matlab (with fspecial) and copy the values. In this case, for a kernel with $t=1$, we store the values in h_kernel. Then we perform the multiplication and compute the inverse Fourier transform. This code excerpt contains the important lines:
DLL int blur(float *h_Data, int dataH, int dataW, float **d, int *dH, int *dW) {
float
*d_Data,
*d_PaddedData,
*d_Kernel,
*d_PaddedKernel;
fComplex
*d_DataSpectrum,
*d_KernelSpectrum;
cufftHandle
fftPlanFwd,
fftPlanInv;
const int kernelH = 7;
const int kernelW = 6;
...
float *h_ResultGPU = (float *)malloc(fftH * fftW * sizeof(float));
cutilSafeCall( cudaMalloc((void **)&d_Data, dataH * dataW * sizeof(float)) );
...
float h_Kernel[] = {0.0001F,0.0006F,0.0016F,0.0016F,0.0006F,0.0001F,
0.0009F,0.0070F,0.0190F,0.0190F,0.0070F,0.0009F,
0.0043F,0.0314F,0.0854F,0.0854F,0.0314F,0.0043F,
0.0070F,0.0518F,0.1407F,0.1407F,0.0518F,0.0070F,
0.0043F,0.0314F,0.0854F,0.0854F,0.0314F,0.0043F,
0.0009F,0.0070F,0.0190F,0.0190F,0.0070F,0.0009F,
0.0001F,0.0006F,0.0016F,0.0016F,0.0006F,0.0001F};
...
cufftSafeCall( cufftExecR2C(fftPlanFwd,
(cufftReal *)d_PaddedKernel, (cufftComplex *)d_KernelSpectrum) );
cutilSafeCall( cutilDeviceSynchronize() );
cufftSafeCall( cufftExecR2C(fftPlanFwd,
(cufftReal *)d_PaddedData, (cufftComplex *)d_DataSpectrum) );
modulateAndNormalize(d_DataSpectrum, d_KernelSpectrum, fftH, fftW, 1);
cufftSafeCall( cufftExecC2R(fftPlanInv,
(cufftComplex *)d_DataSpectrum, (cufftReal *)d_PaddedData) );
cutilSafeCall( cutilDeviceSynchronize() );
cutilSafeCall( cudaMemcpy(h_ResultGPU,
d_PaddedData, fftH * fftW * sizeof(float), cudaMemcpyDeviceToHost) );
...
*d=h_ResultGPU;
*dH=fftH;
*dW=fftW;
return 1;
}
I have highlighted the lines where the Fourier transform, the multiplication and the inverse Fourier transform with the induced blur take place.
In the application, we must have a way to call the DLL function. This code loads the DLL, creates a function pointer and casts it to the DLL function.
private: int CallMyDLL(float *f, int h, int w, float **d, int *dh, int *dw)
{
HINSTANCE hGetProcIDDLL = LoadLibrary(L"cudaint.dll");
FARPROC lpfnGetProcessID =
GetProcAddress(HMODULE (hGetProcIDDLL),"blur");
typedef int (__stdcall * pICFUNC)
(float *, int, int, float**, int *, int*);
pICFUNC blur;
blur = pICFUNC(lpfnGetProcessID);
int MyReturnVal = blur(f, h, w, d, dh, dw);
FreeLibrary(hGetProcIDDLL);
return MyReturnVal;
}
Then we can call with the pixels as a linearized matrix and the computations will be made within the GPU. Conversions between .Net objects and float matrices are straighforward.
$$u_t = \Delta u$$
Where $\Delta$ is the laplacian operator and $u_t$ is the derivative with respect time. The solution at time $t$ is
$$u(x) = \int G_t (x-y) u_o(y) dy$$
Where $x,y\in \mathbb{R}^2$, $u_0$ is the initial noisy image and $G_t$ is the bivariate (normalized) gaussian kernel with standard deviation $t$.
Since a convolution is equivalent to a multiplication of the Fourier transforms as below (by the convolution theorem, and transforming back), we can use the CuFFT library to perform these steps.
$$\hat{u}(\omega) = \hat{G_t}(\omega)\hat{u_o}(\omega)$$
With this project, I want to hunt down two birds with one shot. I want to make a managed code application with Windows forms that reads and visualizes a JPG file, adds noise and then blurs it to remove some noise.
Now, you must know that you cannot mix managed code and CUDA. The solution to this is to separate the CUDA computation and the presentation application in two different parts. We will encapsulate the CUDA computations into a DLL and the Windows program in a Windows Forms application. Both will be part of the same Visual Studio solution so everything will work in the final step. The Windows Forms will be the main project within the VS solution.
Under the DLL project, you must set up CUDA libraries. For the compiler to look for CUDA and CUDAutils header files, we must add
$(NVSDKCOMPUTE_ROOT)\C\common\inc;"$(CUDA_PATH)/include";./;../../common/inc;../../../shared/inc
to project properties -> C/C++ -> General. Then we must add the directory where additional support header files can be found, in this case,
$(NVSDKCOMPUTE_ROOT)\C\common\inc
. This must be done under project properties -> CUDA Runtime API -> Additional Include Directories. In project properties ->Linker -> Additional library directories you must add $(CUDA_PATH)/lib/$(PlatformName);../../common/lib/$(PlatformName);$(NVSDKCOMPUTE_ROOT)\C\common\lib\Win32
Then our DLL function takes a matrix of values and performs the Fourier transform of both (previously padded) the input data and the kernel. We can precompute the kernel with Matlab (with fspecial) and copy the values. In this case, for a kernel with $t=1$, we store the values in h_kernel. Then we perform the multiplication and compute the inverse Fourier transform. This code excerpt contains the important lines:
DLL int blur(float *h_Data, int dataH, int dataW, float **d, int *dH, int *dW) {
float
*d_Data,
*d_PaddedData,
*d_Kernel,
*d_PaddedKernel;
fComplex
*d_DataSpectrum,
*d_KernelSpectrum;
cufftHandle
fftPlanFwd,
fftPlanInv;
const int kernelH = 7;
const int kernelW = 6;
...
float *h_ResultGPU = (float *)malloc(fftH * fftW * sizeof(float));
cutilSafeCall( cudaMalloc((void **)&d_Data, dataH * dataW * sizeof(float)) );
...
float h_Kernel[] = {0.0001F,0.0006F,0.0016F,0.0016F,0.0006F,0.0001F,
0.0009F,0.0070F,0.0190F,0.0190F,0.0070F,0.0009F,
0.0043F,0.0314F,0.0854F,0.0854F,0.0314F,0.0043F,
0.0070F,0.0518F,0.1407F,0.1407F,0.0518F,0.0070F,
0.0043F,0.0314F,0.0854F,0.0854F,0.0314F,0.0043F,
0.0009F,0.0070F,0.0190F,0.0190F,0.0070F,0.0009F,
0.0001F,0.0006F,0.0016F,0.0016F,0.0006F,0.0001F};
...
cufftSafeCall( cufftExecR2C(fftPlanFwd,
(cufftReal *)d_PaddedKernel, (cufftComplex *)d_KernelSpectrum) );
cutilSafeCall( cutilDeviceSynchronize() );
cufftSafeCall( cufftExecR2C(fftPlanFwd,
(cufftReal *)d_PaddedData, (cufftComplex *)d_DataSpectrum) );
modulateAndNormalize(d_DataSpectrum, d_KernelSpectrum, fftH, fftW, 1);
cufftSafeCall( cufftExecC2R(fftPlanInv,
(cufftComplex *)d_DataSpectrum, (cufftReal *)d_PaddedData) );
cutilSafeCall( cutilDeviceSynchronize() );
cutilSafeCall( cudaMemcpy(h_ResultGPU,
d_PaddedData, fftH * fftW * sizeof(float), cudaMemcpyDeviceToHost) );
...
*d=h_ResultGPU;
*dH=fftH;
*dW=fftW;
return 1;
}
I have highlighted the lines where the Fourier transform, the multiplication and the inverse Fourier transform with the induced blur take place.
In the application, we must have a way to call the DLL function. This code loads the DLL, creates a function pointer and casts it to the DLL function.
private: int CallMyDLL(float *f, int h, int w, float **d, int *dh, int *dw)
{
HINSTANCE hGetProcIDDLL = LoadLibrary(L"cudaint.dll");
FARPROC lpfnGetProcessID =
GetProcAddress(HMODULE (hGetProcIDDLL),"blur");
typedef int (__stdcall * pICFUNC)
(float *, int, int, float**, int *, int*);
pICFUNC blur;
blur = pICFUNC(lpfnGetProcessID);
int MyReturnVal = blur(f, h, w, d, dh, dw);
FreeLibrary(hGetProcIDDLL);
return MyReturnVal;
}
Then we can call with the pixels as a linearized matrix and the computations will be made within the GPU. Conversions between .Net objects and float matrices are straighforward.
I made the source code available here.
Monday, April 30, 2012
Miscellaneous errors with Visual C++
If you compile the project and receive the following error:
fatal error LNK1104: cannot open file 'C:\Program.obj'
This particular issue is caused by specifying a dependency to a lib file that had spaces in its path. The path needs to be surrounded by quotes for the project to compile correctly.
On the Configuration Properties -> Linker -> Input tab of the project’s properties, there is an Additional Dependencies property. This issue was fixed by changing this property from:
C:\Program Files\sofware sdk\lib\library.lib
To:
" C:\Program Files\sofware sdk\lib\library.lib"
On the Configuration Properties -> Linker -> Input tab of the project’s properties, there is an Additional Dependencies property. This issue was fixed by changing this property from:
C:\Program Files\sofware sdk\lib\library.lib
To:
" C:\Program Files\sofware sdk\lib\library.lib"
Thursday, April 19, 2012
Shape recognition and shape spaces in computer vision
One of the problems in computer vision and pattern recognition is that of shape recognition (classification, clustering and retrieval). It is important in statistical analysis of medical images (shape outliers might indicate disease) and machine vision (digital recording and analysis based on planar views of 3D objects).
Classical statistical shape analysis is due to Kendall, Bookstein and Mardia. This classical shape analysis treats shapes as shape spaces, whose geometries are those of differentiable manifolds often with appropriate Riemannian structures.
Imagine we are given a set of variables in which each variable is a position of a prearranged landmark (interesting point with a load of information) and our setting is 2D. Then we can model the points as complex numbers $p \in \mathbb{C}$. Therefore each point can be written as $p_i=x_i + iy_i$ where $i=\sqrt{-1}$. A k-ad $p={p_i}_k$ is a set of $k$ shape landmarks.
We now center these points so as to make them translation invariant.
$$w_T=p-\bar{p}$$
where $\bar{p} \in \mathbb{C}^k$ is a vector with all elements set to the k-ad mean, this is $p{_Tj}=\frac{1}{k}\sum_k{p_i}$.
Now we make them scale invariant by dividing each element in $w_T \in \mathbb{C}^k$ by the absolute value of $w_T$, this is $w_S = \frac{w_T}{|w_T|}$.
Finally given the rotation matrix with angle $\theta$
$$\left[\begin{array}{cc}cos(\theta) & -sin(\theta) \\ sin(\theta) & cos(\theta)\end{array}\right]$$
one can rotate the centered shape with $Rw_S$ (if we had modeled it in $\mathbb{R}^{2\times k}$, but we are in the complex domain and it is done by multiplying each element by $e^{i\theta}$. Now, this is when it gets interesting and useful (and where I see the true magic). To make our representation rotation invariant, we can use the Veronese-Whitney embedding, so that the rotation invariant representation is the rank one matrix $w_R=w_S w_S^*$, where $\cdot^*$ is the complex conjugate. This means that, by what we have seen before, any rotation of $w_S$, $e^{i\theta} w_S$ generates
$$e^{i\theta}w_S e^{-i\theta}w_S^* = w_S w_S^*$$
since $e^{-i\theta}$ is the complex conjugate of the rotation (meaning that we undo the rotation by rotating $-\theta$ degrees. This is readily seen by the (complex) exponential arithmetic $e^{i\theta}e^{-i\theta} = = e^{i(\theta - \theta)} = 1$.
This space is that of the complex hermitian matrices and we can define the norm of a shape and the distance between two shapes by
$$\|A\| = trace(A) \\
\rho(A,B) = \|A-B\|$$
Now we can compute the mean shape of populations and do some inference. For example, populations of known healthy livers against livers with disease, and get confidence intervals so as to accept or reject null hypothesis given a new liver shape.
We see that modeling in $\mathbb{R}^{2\times k}$ and using the matrix $R$, we achieve the same result:
$$wR(wR)^T = wRR^T w^T = ww^T$$
Check out: http://stat.duke.edu/~ab216/duke-talk.pdf
Classical statistical shape analysis is due to Kendall, Bookstein and Mardia. This classical shape analysis treats shapes as shape spaces, whose geometries are those of differentiable manifolds often with appropriate Riemannian structures.
Imagine we are given a set of variables in which each variable is a position of a prearranged landmark (interesting point with a load of information) and our setting is 2D. Then we can model the points as complex numbers $p \in \mathbb{C}$. Therefore each point can be written as $p_i=x_i + iy_i$ where $i=\sqrt{-1}$. A k-ad $p={p_i}_k$ is a set of $k$ shape landmarks.
We now center these points so as to make them translation invariant.
$$w_T=p-\bar{p}$$
where $\bar{p} \in \mathbb{C}^k$ is a vector with all elements set to the k-ad mean, this is $p{_Tj}=\frac{1}{k}\sum_k{p_i}$.
Now we make them scale invariant by dividing each element in $w_T \in \mathbb{C}^k$ by the absolute value of $w_T$, this is $w_S = \frac{w_T}{|w_T|}$.
Finally given the rotation matrix with angle $\theta$
$$\left[\begin{array}{cc}cos(\theta) & -sin(\theta) \\ sin(\theta) & cos(\theta)\end{array}\right]$$
one can rotate the centered shape with $Rw_S$ (if we had modeled it in $\mathbb{R}^{2\times k}$, but we are in the complex domain and it is done by multiplying each element by $e^{i\theta}$. Now, this is when it gets interesting and useful (and where I see the true magic). To make our representation rotation invariant, we can use the Veronese-Whitney embedding, so that the rotation invariant representation is the rank one matrix $w_R=w_S w_S^*$, where $\cdot^*$ is the complex conjugate. This means that, by what we have seen before, any rotation of $w_S$, $e^{i\theta} w_S$ generates
$$e^{i\theta}w_S e^{-i\theta}w_S^* = w_S w_S^*$$
since $e^{-i\theta}$ is the complex conjugate of the rotation (meaning that we undo the rotation by rotating $-\theta$ degrees. This is readily seen by the (complex) exponential arithmetic $e^{i\theta}e^{-i\theta} = = e^{i(\theta - \theta)} = 1$.
This space is that of the complex hermitian matrices and we can define the norm of a shape and the distance between two shapes by
$$\|A\| = trace(A) \\
\rho(A,B) = \|A-B\|$$
Now we can compute the mean shape of populations and do some inference. For example, populations of known healthy livers against livers with disease, and get confidence intervals so as to accept or reject null hypothesis given a new liver shape.
We see that modeling in $\mathbb{R}^{2\times k}$ and using the matrix $R$, we achieve the same result:
$$wR(wR)^T = wRR^T w^T = ww^T$$
Check out: http://stat.duke.edu/~ab216/duke-talk.pdf
Setting up CUDA in Windows and Visual C++ 2008, Addendum II
After everything is installed, you will need to set additional options. Credit for this goes to the user Tom of Stackoverflow. I will reproduce here the steps to configure CUDA in newer Toolkit versions. Users interested in older ones should check the link at the end of the post.
In addition to the previous CUDA posts, where you can learn how to install the required tools and make Visual C++ understand what you are using, you need to:
If you have Visual Studio 2010, here are the details for you:
http://stackoverflow.com/questions/3778799/how-do-i-start-a-cuda-app-in-visual-studio-2010/7285235#7285235
For older versions of CUDA:
http://stackoverflow.com/questions/2046228/how-do-i-start-a-new-cuda-project-in-visual-studio-2008
In addition to the previous CUDA posts, where you can learn how to install the required tools and make Visual C++ understand what you are using, you need to:
- Add the
NvCudaRuntimeApi.rulesto the project file (right click on the project, Custom Build Rules, tick the relevant box). - Add the CUDA runtime library (right click on the project and choose Properties, then in Linker -> General add
$(CUDA_PATH)\lib\$(PlatformName)to the Additional Library Directories and in Linker -> Input addcudart.libto the Additional Dependencies). - Optionally add the CUDA include files to the search path, required
if you include any CUDA files in your .cpp files (as opposed to .cu
files) (right click on the project and choose Properties, then in C/C++ -> General add
$(CUDA_PATH)\includeto the Additional Include Directories). - Then just build your project and the .cu files will be compiled to .obj and added to the link automaticaly
- Change the code generation to use statically loaded C runtime to match the CUDA runtime; right click on the project and choose Properties, then in C/C++ -> Code Generation change the Runtime Library to /MT (or /MTd for debug, in which case you will need to mirror this in Runtime API -> Host -> Runtime Library). You might get error LNK4098 otherwise.
LNK4098: defaultlib 'LIBCMT' conflicts with use of other libs) or multiply defined symbols for standard library functions, then this should be your first suspect. Also, you might see LNK2001, LNK2005 when some conflict with the C++ runtime exists. A possible workaround might be using C functions.If you have Visual Studio 2010, here are the details for you:
http://stackoverflow.com/questions/3778799/how-do-i-start-a-cuda-app-in-visual-studio-2010/7285235#7285235
For older versions of CUDA:
http://stackoverflow.com/questions/2046228/how-do-i-start-a-new-cuda-project-in-visual-studio-2008
Monday, April 16, 2012
Setting up CUDA in Windows and Visual C++ Express 2008 Addendum
With CUDA Toolkit 4.x, things are slightly different.
<CUDA_SDK>\C\doc\syntax_highlighting
Syntax highlighting
The necessary files are under<CUDA_SDK>\C\doc\syntax_highlighting
Building
The following procedure is credited to the user sdtougao from NVIDIA forums. This is what you need to build your projects correctly.- Copy the rules files, <NVIDIA GPU Computing Toolkit>\CUDA\v4.0\extras\visual_studio_integration\ rules to the VS2008 installation directory, <Microsoft Visual Studio 9.0>\VC\VCProjectDefaults
- Open a SDK project in VS2008, for example clock.sln, open "custom build rule",select "NvCudaRuntimeApi.rules".
Intellisense
A registry key file is provided so that double-clicking on it adds the necessary keys to the registry without any extra effort.Sunday, April 8, 2012
Setting up CUDA on Windows and Visual C++ Express 2008
Assuming we already have Visual C++ 2008 Express installed, we need to perform the following
steps:
directory.
compiler) and, last, the linking process.
Remark: nVidia compiler, nvcc, treats differently the objects it finds in the .cu files. If it is a kernel, it compiles it into PTX instructions and stores them into a data section of the object file. On the other hand, if it finds a plain C function, it compiles it as any C compiler would do, storing native hardware instructions into the executable text section of the object file. Remark: Since the device code is executed in the GPU, the kernels live as data in the host code and are copied to the device memory in an I/O operation initiated by the code nvcc put when the C function in the .cu file invoked the kernel. This is explained later.
However, if an output like this is experienced:
Command Line” in the Additional Options textbox.
Since the .cu files are compiled by nvcc and kinked to the nVidia libraries (according to the
supplied build rules), a general Visual C++ project links to the standard set of libraries, and
some of them get in conflict with nVidia’s. The previous fix avoids using the conflicting
standard libraries.
steps:
- Download and install the latest nVidia driver with CUDA that is compatible with the system’s graphics card. Then reboot your system so that the driver loads to memory. Note that nVidia publishes a special series of drivers for desktops and for notebooks cards (the later with some time lag).
- Download and install both the CUDA Toolkit and the CUDA SDK. These must be the same version as the driver before. If the examples in the bin directory of the SDK do not work, this indicates you have a wrong version of the driver (and of the whole package).
Syntax highlight and Intellisense
Since the extension .cu is not recognized as C code by Visual C++, it will not highlight the language’s reserved words. NVidia supply a usertype.dat under: “<CUDA SDK install dir>\doc\syntax_highlighting\visual_studio_8”.- Copy this file to: “\Program Files\Microsoft Visual Studio 9.0\Common7\IDE”. Now open Visual Studio and navigate to: tools->options->text editor->file extension. In the extension box type “cu”, make sure you select “Microsoft Visual C++”.
- On Windows 7, open the registry editor and navigate to “HKEY_USERS\<user code>\Software\Microsoft\VCExpress\9.0\Languages\Language Services\C/C++” then go to the string entry “NCB Default C/C++ Extensions” and add “;.cu;.cuh” to the end of the list. The <user code> is machine-specific, the best thing to do is look for “NCB Default C/C++ Extensions” in the registry editor. Restart Visual Studio and Intellisense will work for your CUDA files. If there is a specific CUDA C feature, it will not recognize it, but all the normal C keywords will be benefited.
Building
nVidia supply a rules file which you will find under the “<CUDA SDK install dir>\common”directory.
- Select it under “project -> custom build rules -> find existing”. This will allow the IDE to find the compiler nvcc for the .cu files.
Debug Win32 ------Notice that the compilation of the .cu files (with nvcc), then the .cpp files (with the Visual C++
1>Compiling with CUDA Build Rule...
1>"D:\my\CUDA\bin\nvcc.exe" -arch sm_10 -ccbin "C:\Program Files\Microsoft Visual Studio
9.0\VC\bin" -use_fast_math -Xcompiler "/EHsc /W3 /nologo /O2 /Zi /MT " -
I"D:\my\CUDA_SDK\common\inc" -maxrregcount=32 --compile -o "Debug\kernel.cu.obj"
"c:\Users\Gabo\Documents\Visual Studio 2008\Projects\more\cuda\kernel.cu"
1>kernel.cu
1>tmpxft_00001314_00000000-3_kernel.cudafe1.gpu
1>tmpxft_00001314_00000000-8_kernel.cudafe2.gpu
1>tmpxft_00001314_00000000-3_kernel.cudafe1.cpp
1>tmpxft_00001314_00000000-13_kernel.ii
1>Compilando...
1>main.cpp
1>math.cpp
1>Building code...
1>Linking...
1>main.obj : warning LNK4075: se omite '/EDITANDCONTINUE' due to the specification
'/INCREMENTAL:NO'
1>Incrustando manifiesto...
1>El registro de compilación se guardó en el "file://c:\Users\Gabo\Documents\Visual Studio
2008\Projects\more\cuda\Debug\BuildLog.htm"
1>cuda - 0 errors, 3 warnings
compiler) and, last, the linking process.
Remark: nVidia compiler, nvcc, treats differently the objects it finds in the .cu files. If it is a kernel, it compiles it into PTX instructions and stores them into a data section of the object file. On the other hand, if it finds a plain C function, it compiles it as any C compiler would do, storing native hardware instructions into the executable text section of the object file. Remark: Since the device code is executed in the GPU, the kernels live as data in the host code and are copied to the device memory in an I/O operation initiated by the code nvcc put when the C function in the .cu file invoked the kernel. This is explained later.
Linking difficulties
For a seamless compilation, the best thing to do is to copy an existing CUDA project from the SDK to a separate folder, including the libraries (from the “common” directory), rename and redistribute everything and start from thereHowever, if an output like this is experienced:
1>MSVCRTD.lib(MSVCR90D.dll) : error LNK2005: ya se definió _free en LIBCMT.lib(free.obj)Adding /NODEFAULTLIB:LIBCMT.lib to “project -> Configuration Properties -> Linker ->
1>MSVCRTD.lib(MSVCR90D.dll) : error LNK2005: ya se definió _malloc en LIBCMT.lib(malloc.obj)
1>MSVCRTD.lib(MSVCR90D.dll) : error LNK2005: ya se definió _printf en LIBCMT.lib(printf.obj)
1>MSVCRTD.lib(ti_inst.obj) : error LNK2005: ya se definió "private: __thiscall type_info::type_info(class type_info const &)" (??0type_info@@AAE@ABV0@@Z) en LIBCMT.lib(typinfo.obj)
1>MSVCRTD.lib(ti_inst.obj) : error LNK2005: ya se definió "private: class type_info & __thiscall type_info::operator=(class type_info const &)" (??4type_info@@AAEAAV0@ABV0@@Z) en LIBCMT.lib(typinfo.obj)
Command Line” in the Additional Options textbox.
Since the .cu files are compiled by nvcc and kinked to the nVidia libraries (according to the
supplied build rules), a general Visual C++ project links to the standard set of libraries, and
some of them get in conflict with nVidia’s. The previous fix avoids using the conflicting
standard libraries.
Subscribe to:
Posts (Atom)