Intel SPMD (ISPC) with CMake

####################
# add it to your CMakeLists.txt
#################### ISPC

# given an ispc file, append it and the generated .h file to 'outList'
macro( AddIspcSrc outList outListObj srcLocationsList )
  file(MAKE_DIRECTORY ${CMAKE_CURRENT_BINARY_DIR}/${CMAKE_CFG_INTDIR}/CMakeFiles/ispc)

  foreach(ispc_file ${srcLocationsList})
    message("ISPC source file " ${ispc_file})
    set(srcLocation2 ${ispc_file})
    find_file(srcLocation2 NAMES ${ispc_file} PATHS ${CMAKE_CURRENT_SOURCE_DIR} )
    get_source_file_property(srcLocation2 ${srcLocation2} LOCATION)
    string(FIND "${ispc_file}" "/" lastSlash REVERSE) #just get the file name, strip the path
    if(NOT ${lastSlash} STREQUAL "-1" )
      math(EXPR lastSlash ${lastSlash}+1)
      string(SUBSTRING "${ispc_file}" ${lastSlash} -1 objLocation)
    else()
      set( objLocation ${ispc_file} )
    endif()
    string(REPLACE ".ispc" ".obj" objLocation ${objLocation})
    set(objLocation ${CMAKE_CURRENT_BINARY_DIR}/${CMAKE_CFG_INTDIR}/CMakeFiles/ispc/${objLocation}) # put the obj in the intermediate dir
    string(REPLACE ".ispc" "_ispc.h" hdrLocation ${ispc_file})

    message("ISPC header " ${hdrLocation})
    message("ISPC object" ${objLocation})
    ADD_CUSTOM_COMMAND(
       OUTPUT ${objLocation}
       COMMAND ispc ${srcLocation2} -o ${objLocation} -h ${hdrLocation} --arch=x86-64 --target=avx1-i32x8
       IMPLICIT_DEPENDS C ${ispc_file}
       DEPENDS ${ispc_file}
       MAIN_DEPENDENCY  ${ispc_file}
       )
    set_source_files_properties( ${objLocation} PROPERTIES GENERATED TRUE EXTERNAL_OBJECT TRUE )
    set_source_files_properties( ${hdrLocation} PROPERTIES GENERATED TRUE EXTERNAL_OBJECT TRUE )
    list( APPEND ${outList} ${srcLocation} ${hdrLocation} )
    list( APPEND ${outListObj} ${objLocation})
  endforeach(ispc_file)
endmacro( AddIspcSrc )

# collect ispc files
file(GLOB_RECURSE ISPC_FILES  ${PROJECT_SOURCE_DIR}/*.ispc)

# call ispc preprocessor, collect *obj files in outListObj
AddIspcSrc( outList outListObj "${ISPC_FILES}")

add_custom_target(ISPC_OBJECTS ALL DEPENDS ${outListObj})


# your exe file
add_executable(${PROJECT_NAME} ${CPP_FILES})

# make a library of all ispc *obj file
add_library( ispc_objs STATIC EXCLUDE_FROM_ALL ${outListObj})
add_dependencies(ispc_objs ISPC_OBJECTS)
set_target_properties(ispc_objs PROPERTIES LINKER_LANGUAGE C)


# link ispc-ed library with your exe
target_link_libraries(${PROJECT_NAME} ispc_objs)

Unit-testing for CUDA with Google C++ Testing Framework

This code is supposed to be situated in /tests folder as a subdirectory of CMake-based building system

##########################################
##########################################
##########################################
##########################################
##########################################
##########################################
###                     CMakeLists.txt


cmake_minimum_required(VERSION 2.8.7 FATAL_ERROR)
project(tests)

# enable ExternalProject CMake module
include(ExternalProject)

# Set default ExternalProject root directory
set_directory_properties(PROPERTIES EP_PREFIX ${CMAKE_BINARY_DIR}/ThirdParty)

ExternalProject_Add(
    googletest
    URL http://googletest.googlecode.com/files/gtest-1.6.0.zip
    # TIMEOUT 10
    # # Force separate output paths for debug and release builds to allow easy
    # # identification of correct lib in subsequent TARGET_LINK_LIBRARIES commands
    # CMAKE_ARGS -DCMAKE_ARCHIVE_OUTPUT_DIRECTORY_DEBUG:PATH=DebugLibs
    #            -DCMAKE_ARCHIVE_OUTPUT_DIRECTORY_RELEASE:PATH=ReleaseLibs
    #            -Dgtest_force_shared_crt=ON
    # Disable install step
    INSTALL_COMMAND ""
    # Wrap download, configure and build steps in a sсript to log output
    LOG_DOWNLOAD ON
    LOG_CONFIGURE ON
    LOG_BUILD ON)
# Specify include dir
ExternalProject_Get_Property(googletest source_dir)
set(GTEST_INCLUDE_DIR ${source_dir}/include)

# Library
ExternalProject_Get_Property(googletest binary_dir)
set(GTEST_LIBRARY_PATH ${binary_dir}/${CMAKE_FIND_LIBRARY_PREFIXES}gtest.a)
set(GTEST_LIBRARY gtest)

add_library(${GTEST_LIBRARY} UNKNOWN IMPORTED)
set_property(TARGET ${GTEST_LIBRARY}
             PROPERTY IMPORTED_LOCATION ${GTEST_LIBRARY_PATH} )

add_dependencies(${GTEST_LIBRARY} googletest)


# Add test executable target
add_executable(maintest ${PROJECT_SOURCE_DIR}/test.cpp)
# Create dependency of MainTest on googletest
add_dependencies(maintest googletest)

cuda_add_library(cudatests ${GPU_LIBRARY_TYPE}
${PROJECT_SOURCE_DIR}/kernels.cu
OPTIONS ${CUDAOPTIONS}
)

# Specify MainTests link libraries
ExternalProject_Get_Property(googletest binary_dir)

target_link_libraries(maintest ${GTEST_LIBRARY} cudatests)

////////////////////////////////////////////
////////////////////////////////////////////
////////////////////////////////////////////
//                            kernels.cu

#include "kernels.cuh"

__global__
void
square_array_kernel(float* a, unsigned int numElements)
{
  // ### implement me ###
  int i = blockIdx.x * blockDim.x + threadIdx.x;

  if (i < numElements)
    a[i] = a[i] * a[i];
}

void
test()
{
  float*             a_host;                // pointer to array in host memory
  const unsigned int numElements = 100000000;      // number of elements in the array
  size_t             size        = numElements * sizeof(float);
  a_host = new float[size];                 // allocate array on host

  float* a_device;

  cudaMalloc((void**)&a_device, size);
  cudaMemcpy(a_device, a_host, size, cudaMemcpyHostToDevice);

  int block_size = 256;
  int grid_size  = (numElements + block_size - 1) / block_size;

  square_array_kernel << < grid_size, block_size >> > (a_device, numElements);

  cudaMemcpy(a_host, a_device, size, cudaMemcpyDeviceToHost);
  cudaFree(a_device);

  delete[] a_host;
}

////////////////////////////////////////////
////////////////////////////////////////////
////////////////////////////////////////////
//                            kernels.cuh


#ifndef _KERNELS_H_
#define _KERNELS_H_

__global__
void
square_array_kernel(float* a, unsigned int numElements);

void
test();

#endif //  _KERNELS_H_


////////////////////////////////////////////
////////////////////////////////////////////
////////////////////////////////////////////
//                            test.cpp 

#include "gtest/gtest.h"

#include <cuda.h>
#include <cuda_runtime.h>

#include "kernels.cuh"

TEST(A, B) {
  test();

  SUCCEED();
}

int
main(int argc, char** argv)
{
  testing::InitGoogleTest(&argc, argv);
  return RUN_ALL_TESTS();
}

Resize images with Python

Recently, I needed to resize some photos in order to upload them to the webpage. There was no point in uploading the full-sized photos, so I downscaled them with a ratio 1/3. The script should be started from the directory with your JPEGs. The output is a sequence of files: resized_0.jpg, resized_1.jpg, resized_2.jpg etc

from os import walk
import Image

def imgResize(im):
    div = 3
    width = im.size[0] / div
    height = im.size[1] / div

    im_result = im.resize((width, height), Image.ANTIALIAS)
    # best down-sizing filter

    return im_result


files = []
for (dirpath, dirnames, filenames) in walk("."):
    files.extend(filenames)
    break

for i, f in enumerate(files):
    if "JPG" in f and not "resized" in f:
        im = Image.open(f)
        im = imgResize(im)
        im.save("resized_"+str(i)+".jpg")

Многомерная сплайн-интерполяция. Построение карт

Моя старая магистерская диссертация на тему многомерной интерполяции данных в геофизике.

https://docs.google.com/file/d/0B32kbaZOrKltT0NJSWphUHJySnM/edit