Create a gist now

Instantly share code, notes, and snippets.

CUDA Jpeg2000 Windows port
Index: gpu_jpeg2k/CMake/FindFreeImage.cmake
===================================================================
--- gpu_jpeg2k/CMake/FindFreeImage.cmake (revision 84)
+++ gpu_jpeg2k/CMake/FindFreeImage.cmake (working copy)
@@ -14,7 +14,7 @@
FIND_LIBRARY( FREEIMAGE_LIBRARY
NAMES FreeImage freeimage
PATHS
- ${PROJECT_SOURCE_DIR}/FreeImage
+ ${PROJECT_SOURCE_DIR}/extern/FreeImage
DOC "The FreeImage library")
ELSE (WIN32)
FIND_PATH( FREEIMAGE_INCLUDE_PATH FreeImage.h
Index: gpu_jpeg2k/CMakeLists.txt
===================================================================
--- gpu_jpeg2k/CMakeLists.txt (revision 84)
+++ gpu_jpeg2k/CMakeLists.txt (working copy)
@@ -6,8 +6,15 @@
#Uncomment next line to enable debugging (for use with cuda-gdb)
#SET(CUDA_NVCC_FLAGS -g -G --compiler-options -fpermissive)
+IF(UNIX)
SET(CMAKE_CXX_FLAGS "-O3 -fno-builtin")
SET(CMAKE_CC_FLAGS "-O3 -fno-builtin")
+ENDIF(UNIX)
+IF(MSVC)
+SET(CMAKE_CXX_FLAGS "/Ox")
+SET(CMAKE_CC_FLAGS "/Ox")
+SET(CMAKE_EXE_LINKER_FLAGS "/FORCE:MULTIPLE")
+ENDIF(MSVC)
SET(CUDA_NVCC_FLAGS -O3 --compiler-options -fpermissive -arch sm_20 --maxrregcount 64 --ptxas-options=-v)
#SET(CUDA_NVCC_FLAGS --compiler-options -fpermissive -arch sm_20 --maxrregcount 64 --ptxas-options=-v)
SET(CUDA_VERBOSE_BUILD ON)
@@ -36,7 +43,6 @@
TARGET_LINK_LIBRARIES( encoder
${FREEIMAGE_LIBRARY}
-m
gpu_coeff_coder
#ebcot
config
@@ -55,7 +61,6 @@
TARGET_LINK_LIBRARIES( decoder
${FREEIMAGE_LIBRARY}
-m
gpu_coeff_coder
config
dwt
Index: gpu_jpeg2k/config/CMakeLists.txt
===================================================================
--- gpu_jpeg2k/config/CMakeLists.txt (revision 84)
+++ gpu_jpeg2k/config/CMakeLists.txt (working copy)
@@ -2,4 +2,5 @@
parameters.c
ini.c
arguments.c
-init_device.c)
+init_device.c
+getopt.c)
Index: gpu_jpeg2k/config/arguments.c
===================================================================
--- gpu_jpeg2k/config/arguments.c (revision 84)
+++ gpu_jpeg2k/config/arguments.c (working copy)
@@ -26,7 +26,9 @@
#include <string.h>
#include <stdio.h>
#include <stdlib.h>
-#include <unistd.h>
+//#include <unistd.h>
+#include "getopt.h"
+#include <io.h>
#include "arguments.h"
#include "../types/image_types.h"
@@ -36,6 +38,7 @@
int parse_args(int argc, char **argv, type_image *img) {
uint8_t flag = 1;
char c;
+ int i = 0;
while ((c = getopt(argc, argv, "i:o:h:c:")) != -1)
switch (c) {
case 'i':
@@ -78,8 +81,6 @@
default:
return ERROR;
}
-
- int i = 0;
for (i = optind; i < argc; ++i) {
printf("Non-option argument %s\n", argv[i]);
}
Index: gpu_jpeg2k/config/init_device.c
===================================================================
--- gpu_jpeg2k/config/init_device.c (revision 84)
+++ gpu_jpeg2k/config/init_device.c (working copy)
@@ -31,10 +31,11 @@
{
int *d_tmp;
int h_tmp = 1;
+ size_t printf_buff;
cuda_set_device_flags();
- size_t printf_buff = 10 * 1024 * 1024 * 100;
+ printf_buff = 10 * 1024 * 1024 * 100;
cuda_set_printf_limit(printf_buff);
cuda_set_device(param->param_device);
Index: gpu_jpeg2k/decoder.c
===================================================================
--- gpu_jpeg2k/decoder.c (revision 84)
+++ gpu_jpeg2k/decoder.c (working copy)
@@ -61,6 +61,11 @@
int main(int argc, char **argv)
{
// println_start(INFO);
+ type_parameters *param;
+ FILE *fsrc;
+ type_tile *tile;
+ int i;
+
type_image *img = (type_image *)malloc(sizeof(type_image));
memset(img, 0, sizeof(type_image));
if((parse_args(argc, argv, img) == ERROR) || (check_args_dec(img) == ERROR))
@@ -70,7 +75,7 @@
return 1;
}
- type_parameters *param = (type_parameters*)malloc(sizeof(type_parameters));
+ param = (type_parameters*)malloc(sizeof(type_parameters));
if((parse_config(img->conf_file, param) == ERROR) || (check_config(param) == ERROR)) {
fprintf(stderr, "Error occurred while parsing configuration file.\n");
fprintf(stdout, "%s", help);
@@ -79,15 +84,12 @@
init_device(param);
- FILE *fsrc = fopen(img->in_file, "rb");
+ fsrc = fopen(img->in_file, "rb");
if (!fsrc) {
fprintf(stderr, "Error, failed to open %s for reading\n", img->in_file);
return 1;
}
- type_tile *tile;
- int i;
-
if(strstr(img->in_file, ".jp2") != NULL) {
println(INFO, "It's a JP2 file");
Index: gpu_jpeg2k/dwt/fwt.cu
===================================================================
--- gpu_jpeg2k/dwt/fwt.cu (revision 84)
+++ gpu_jpeg2k/dwt/fwt.cu (working copy)
@@ -174,6 +174,10 @@
void fprocess_97(short tidy, const short tidx2, short p_offset_y, float *pix_neighborhood, const float shared[][MEMSIZE + 1],
float *results)
{
+ const float a1 = -1.586134342f;
+ const float a2 = -0.05298011854f;
+ const float a3 = 0.8829110762f;
+ const float a4 = 0.4435068522f;
// Read necessary data
#pragma unroll
for (int i = 0; i < 9; i++)
@@ -248,6 +252,26 @@
void fprocess_53(short tidy, const short tidx2, short p_offset_y, int *pix_neighborhood, const int shared[][MEMSIZE + 1],
int *results)
{
+ /**
+ * @defgroup ScaleCoeff Scale coefficients.
+ *
+ * Scale coefficients.
+ *
+ * @{
+ */
+const float k = 1.230174104914f; // 1.230174104914
+/** @} */
+
+/**
+ * @defgroup 53Coeff 53 coefficients.
+ *
+ * 53 coefficients.
+ *
+ * @{
+ */
+const float p53 = -0.5f;
+const float u53 = 0.25f;
+/** @} */
// Read necessary data
#pragma unroll
for (int i = 0; i < 5; i++)
@@ -407,6 +431,26 @@
__global__
void fwt97(const float *idata, float *odata, const int2 img_size, const int2 step)
{
+ /**
+ * @defgroup ScaleCoeff Scale coefficients.
+ *
+ * Scale coefficients.
+ *
+ * @{
+ */
+const float k = 1.230174104914f; // 1.230174104914
+/** @} */
+
+/**
+ * @defgroup 53Coeff 53 coefficients.
+ *
+ * 53 coefficients.
+ *
+ * @{
+ */
+const float p53 = -0.5f;
+const float u53 = 0.25f;
+/** @} */
/* Shared memory for part of the signal */
__shared__ float shared[MEMSIZE][MEMSIZE + 1];
Index: gpu_jpeg2k/dwt/fwt_1d.cu
===================================================================
--- gpu_jpeg2k/dwt/fwt_1d.cu (revision 84)
+++ gpu_jpeg2k/dwt/fwt_1d.cu (working copy)
@@ -73,6 +73,28 @@
__device__ static void save_data(int tidx, type_data **data, int w, float *pix_neighborhood, int num_components)
{
+ /**
+ * @defgroup 97Coeff 97 Coefficients.
+ *
+ * 97 Coefficients.
+ *
+ * @{
+ */
+const float a1 = -1.586134342f;
+const float a2 = -0.05298011854f;
+const float a3 = 0.8829110762f;
+const float a4 = 0.4435068522f;
+/** @} */
+
+/**
+ * @defgroup ScaleCoeff Scale coefficients.
+ *
+ * Scale coefficients.
+ *
+ * @{
+ */
+const float k = 1.230174104914f; // 1.230174104914
+/** @} */
int pix = blockIdx.x + w * blockIdx.y;
int high_pass = (num_components + 1) >> 1;
@@ -107,6 +129,28 @@
__device__ void process_97(int tidx, float *pix_neighborhood, float *sm, int offset)
{
+ /**
+ * @defgroup 97Coeff 97 Coefficients.
+ *
+ * 97 Coefficients.
+ *
+ * @{
+ */
+const float a1 = -1.586134342f;
+const float a2 = -0.05298011854f;
+const float a3 = 0.8829110762f;
+const float a4 = 0.4435068522f;
+/** @} */
+
+/**
+ * @defgroup ScaleCoeff Scale coefficients.
+ *
+ * Scale coefficients.
+ *
+ * @{
+ */
+const float k = 1.230174104914f; // 1.230174104914
+/** @} */
// Read necessary data
#pragma unroll
for (int i = 0; i < 9; i++)
Index: gpu_jpeg2k/dwt/fwt_new.cu
===================================================================
--- gpu_jpeg2k/dwt/fwt_new.cu (revision 84)
+++ gpu_jpeg2k/dwt/fwt_new.cu (working copy)
@@ -179,6 +179,10 @@
void fprocess_97_new(short tidy, const short tidx2, short p_offset_y, float *pix_neighborhood, const float shared[][MEMSIZE + 1],
float *results)
{
+ const float a1 = -1.586134342f;
+const float a2 = -0.05298011854f;
+const float a3 = 0.8829110762f;
+const float a4 = 0.4435068522f;
// Read necessary data
#pragma unroll
for (int i = 0; i < 9; i++)
@@ -253,6 +257,26 @@
void fprocess_53_new(short tidy, const short tidx2, short p_offset_y, int *pix_neighborhood, const int shared[][MEMSIZE + 1],
int *results)
{
+ /**
+ * @defgroup ScaleCoeff Scale coefficients.
+ *
+ * Scale coefficients.
+ *
+ * @{
+ */
+const float k = 1.230174104914f; // 1.230174104914
+/** @} */
+
+/**
+ * @defgroup 53Coeff 53 coefficients.
+ *
+ * 53 coefficients.
+ *
+ * @{
+ */
+const float p53 = -0.5f;
+const float u53 = 0.25f;
+/** @} */
// Read necessary data
#pragma unroll
for (int i = 0; i < 5; i++)
@@ -417,6 +441,30 @@
__global__
void fwt97_new(const float *idata, float *odata, const int2 img_size, const int2 step)
{
+ const float a1 = -1.586134342f;
+ const float a2 = -0.05298011854f;
+ const float a3 = 0.8829110762f;
+ const float a4 = 0.4435068522f;
+ /**
+ * @defgroup ScaleCoeff Scale coefficients.
+ *
+ * Scale coefficients.
+ *
+ * @{
+ */
+const float k = 1.230174104914f; // 1.230174104914
+/** @} */
+
+/**
+ * @defgroup 53Coeff 53 coefficients.
+ *
+ * 53 coefficients.
+ *
+ * @{
+ */
+const float p53 = -0.5f;
+const float u53 = 0.25f;
+/** @} */
/* Shared memory for part of the signal */
__shared__ float shared[MEMSIZE][MEMSIZE + 1];
Index: gpu_jpeg2k/dwt/iwt.cu
===================================================================
--- gpu_jpeg2k/dwt/iwt.cu (revision 84)
+++ gpu_jpeg2k/dwt/iwt.cu (working copy)
@@ -152,6 +152,28 @@
void iprocess_97(const short tidx2, short tidy, short p_offset_y, const short p_size_x, const short p_size_y,
float *pix_neighborhood, const float shared[][sm_size], float *results)
{
+ /**
+ * @defgroup 97Coeff 97 Coefficients.
+ *
+ * 97 Coefficients.
+ *
+ * @{
+ */
+const float a1 = -1.586134342f;
+const float a2 = -0.05298011854f;
+const float a3 = 0.8829110762f;
+const float a4 = 0.4435068522f;
+/** @} */
+
+/**
+ * @defgroup ScaleCoeff Scale coefficients.
+ *
+ * Scale coefficients.
+ *
+ * @{
+ */
+const float k = 1.230174104914f; // 1.230174104914
+/** @} */
// Process rows
while (tidy < p_size_y && tidx2 < p_size_x)
{
@@ -556,6 +578,28 @@
__global__
void iwt97(const float *idata, float *odata, const int2 img_size, const int2 step)
{
+ /**
+ * @defgroup 97Coeff 97 Coefficients.
+ *
+ * 97 Coefficients.
+ *
+ * @{
+ */
+const float a1 = -1.586134342f;
+const float a2 = -0.05298011854f;
+const float a3 = 0.8829110762f;
+const float a4 = 0.4435068522f;
+/** @} */
+
+/**
+ * @defgroup ScaleCoeff Scale coefficients.
+ *
+ * Scale coefficients.
+ *
+ * @{
+ */
+const float k = 1.230174104914f; // 1.230174104914
+/** @} */
// Shared memory for part of the signal
__shared__ float shared[MEMSIZE + 4][MEMSIZE + 4 + 1];
Index: gpu_jpeg2k/dwt/iwt_1d.cu
===================================================================
--- gpu_jpeg2k/dwt/iwt_1d.cu (revision 84)
+++ gpu_jpeg2k/dwt/iwt_1d.cu (working copy)
@@ -61,6 +61,28 @@
__device__ static void read_data(float *sm, int tidx, type_data** data, int w, int h, int num_components, short offset)
{
+ /**
+ * @defgroup 97Coeff 97 Coefficients.
+ *
+ * 97 Coefficients.
+ *
+ * @{
+ */
+const float a1 = -1.586134342f;
+const float a2 = -0.05298011854f;
+const float a3 = 0.8829110762f;
+const float a4 = 0.4435068522f;
+/** @} */
+
+/**
+ * @defgroup ScaleCoeff Scale coefficients.
+ *
+ * Scale coefficients.
+ *
+ * @{
+ */
+const float k = 1.230174104914f; // 1.230174104914
+/** @} */
const short p_offset_lowpass_l = ((tidx < offset) ? (offset - tidx) /* left symmetric extension*/: -offset + tidx /* take normally pixels */);
const short p_offset_highpass_l = ((tidx < offset) ? (offset - 1 - tidx) /* left symmetric extension*/: -offset + tidx /* take normally pixels */);
int p_offset_lowpass;
@@ -121,6 +143,28 @@
__device__ void iprocess_97(int tidx, float *pix_neighborhood, float *sm, int offset)
{
+ /**
+ * @defgroup 97Coeff 97 Coefficients.
+ *
+ * 97 Coefficients.
+ *
+ * @{
+ */
+const float a1 = -1.586134342f;
+const float a2 = -0.05298011854f;
+const float a3 = 0.8829110762f;
+const float a4 = 0.4435068522f;
+/** @} */
+
+/**
+ * @defgroup ScaleCoeff Scale coefficients.
+ *
+ * Scale coefficients.
+ *
+ * @{
+ */
+const float k = 1.230174104914f; // 1.230174104914
+/** @} */
// Read necessary data
#pragma unroll
for (int i = 0; i < 10; i++)
Index: gpu_jpeg2k/dwt/iwt_new.cu
===================================================================
--- gpu_jpeg2k/dwt/iwt_new.cu (revision 84)
+++ gpu_jpeg2k/dwt/iwt_new.cu (working copy)
@@ -152,6 +152,28 @@
void iprocess_97_new(const short tidx2, short tidy, short p_offset_y, const short p_size_x, const short p_size_y,
float *pix_neighborhood, const float shared[][sm_size], float *results)
{
+ /**
+ * @defgroup 97Coeff 97 Coefficients.
+ *
+ * 97 Coefficients.
+ *
+ * @{
+ */
+const float a1 = -1.586134342f;
+const float a2 = -0.05298011854f;
+const float a3 = 0.8829110762f;
+const float a4 = 0.4435068522f;
+/** @} */
+
+/**
+ * @defgroup ScaleCoeff Scale coefficients.
+ *
+ * Scale coefficients.
+ *
+ * @{
+ */
+const float k = 1.230174104914f; // 1.230174104914
+/** @} */
// Process rows
while (tidy < p_size_y && tidx2 < p_size_x)
{
@@ -453,6 +475,28 @@
__global__
void iwt97_new(const float *idata, float *odata, const int2 img_size, const int2 step)
{
+ /**
+ * @defgroup 97Coeff 97 Coefficients.
+ *
+ * 97 Coefficients.
+ *
+ * @{
+ */
+const float a1 = -1.586134342f;
+const float a2 = -0.05298011854f;
+const float a3 = 0.8829110762f;
+const float a4 = 0.4435068522f;
+/** @} */
+
+/**
+ * @defgroup ScaleCoeff Scale coefficients.
+ *
+ * Scale coefficients.
+ *
+ * @{
+ */
+const float k = 1.230174104914f; // 1.230174104914
+/** @} */
// Shared memory for part of the signal
__shared__ float shared[MEMSIZE][MEMSIZE + 1];
Index: gpu_jpeg2k/encoder.c
===================================================================
--- gpu_jpeg2k/encoder.c (revision 84)
+++ gpu_jpeg2k/encoder.c (working copy)
@@ -28,7 +28,9 @@
#include <stdlib.h>
#include <stdio.h>
+#ifndef _WIN32
#include <unistd.h>
+#endif
#include <string.h>
#include "print_info/print_info.h"
@@ -71,6 +73,11 @@
int main(int argc, char **argv)
{
// println_start(INFO);
+ type_parameters *param;
+ int i = 0;
+ type_tile *tile = NULL;
+ long int start_global;
+
type_image *img = (type_image *)malloc(sizeof(type_image));
memset(img, 0, sizeof(type_image));
if((parse_args(argc, argv, img) == ERROR) || (check_args_enc(img) == ERROR))
@@ -80,7 +87,7 @@
return 1;
}
- type_parameters *param = (type_parameters*)malloc(sizeof(type_parameters));
+ param = (type_parameters*)malloc(sizeof(type_parameters));
if((parse_config(img->conf_file, param) == ERROR) || (check_config(param) == ERROR)) {
fprintf(stderr, "Error occurred while parsing configuration file.\n");
fprintf(stdout, "%s", help);
@@ -105,7 +112,6 @@
#endif
#ifdef GLOBAL_TIME
- long int start_global;
start_global = start_measure();
#endif
@@ -120,8 +126,6 @@
cudaThreadSynchronize();
printf("%ld\n", stop_measure(start_mct));
#endif
- int i = 0;
- type_tile *tile = NULL;
// Do processing for all tiles
for(i = 0; i < img->num_tiles; i++)
{
Index: gpu_jpeg2k/file_format/boxes.c
===================================================================
--- gpu_jpeg2k/file_format/boxes.c (revision 84)
+++ gpu_jpeg2k/file_format/boxes.c (working copy)
@@ -40,10 +40,9 @@
//dest has to be n+1 long
char *sstrncpy(char *dest, const char *src, size_t n) {
+ size_t i;
println_start(INFO);
- size_t i;
-
for (i = 0; i < n; i++) {
// printf("i: %i\n", i);
dest[i] = src[i];
@@ -91,11 +90,11 @@
//box *get_next_box(unsigned char *mem, long int *pos) {
box *get_next_box(FILE *fd) {
+ int read = 0;
+ box *box = init_box();
println_start(INFO);
//box *box = malloc(sizeof(box));
- int read = 0;
- box *box = init_box();
if( (box->lbox = read_bytes(box->lbox, fd, 4)) == NULL)
return NULL;
@@ -126,13 +125,17 @@
}
box *get_next_box_char(box *superbox) {
+ char *content;
+ box *box;
+ int read;
+ int size;
if(superbox->read >= superbox->content_length)
return NULL;
- char *content = superbox->dbox;
- box *box = malloc(sizeof(box));
+ content = superbox->dbox;
+ box = malloc(sizeof(box));
- int read = superbox->read;
+ read = superbox->read;
box->lbox = malloc(5 * sizeof(char));
box->lbox = sstrncpy(box->lbox, &content[read], 4); read += 4;
box->length = hex_to_long(box->lbox, 4);
@@ -151,7 +154,7 @@
read += 8;
}
- int size = (box->length - (read - superbox->read));
+ size = (box->length - (read - superbox->read));
box->dbox = malloc((size+1) * sizeof(char));
box->dbox = sstrncpy(box->dbox, &content[read], size);
read += size;
@@ -187,6 +190,10 @@
}
int h_filetype_box(box *b, type_image *img) {
+ char *minv;
+ int left;
+ char *cl;
+ int i = 1;
char *br = malloc(5 * sizeof(char));
br = strncpy(br, b->dbox, 4);
@@ -196,7 +203,7 @@
} else
println(INFO, "Conforms to IS 15444-1");
- char *minv = malloc(5 * sizeof(char));
+ minv = malloc(5 * sizeof(char));
minv = strncpy(minv, &(b->dbox[4]), 4);
if(hex_to_long(minv, 4) != 0) {
@@ -204,10 +211,9 @@
} else
println(INFO, "MinV OK");
- int left = b->content_length - 8; //16 bytes already read from the box's contents: br,minv
+ left = b->content_length - 8; //16 bytes already read from the box's contents: br,minv
printf("left: %i\n", left);
- char *cl;
- int i = 1;
+
while(left) {
cl = malloc(5 * sizeof(char));
cl = strncpy(cl, &(b->dbox)[4 + i*4], 4); left -= 4; i++;
@@ -228,15 +234,17 @@
}
int h_image_header_box(box *b, type_image *img) {
+ char *cwidth;
+ char *cnum_comp;
char *cheight = malloc(5 * sizeof(char));
cheight = strncpy(cheight, b->dbox, 4);
img->height = hex_to_long(cheight, 4);
- char *cwidth = malloc(5 * sizeof(char));
+ cwidth = malloc(5 * sizeof(char));
cwidth = strncpy(cwidth, &(b->dbox)[4], 4);
img->width = hex_to_long(cwidth, 4);
- char *cnum_comp = malloc(3 * sizeof(char));
+ cnum_comp = malloc(3 * sizeof(char));
cnum_comp = strncpy(cnum_comp, &(b->dbox)[8], 2);
img->num_components = hex_to_long(cnum_comp, 2);
@@ -255,9 +263,11 @@
}
int h_header_box(box *header_box, type_image *img) {
+ box *ihdr;
+ box *b;
println_start(INFO);
header_box->read = 0;
- box *ihdr = get_next_box_char(header_box); //TODO: extract box from within b
+ ihdr = get_next_box_char(header_box); //TODO: extract box from within b
if(hex_to_long(ihdr->tbox, 4) != IMAGE_HEADER_BOX) {
println(INFO, "Image Header Box should be the first one in Header superbox. Exitting!");
@@ -265,7 +275,7 @@
} else
h_image_header_box(ihdr, img);
- box *b;
+
while( (b = get_next_box_char(header_box)) != NULL ) {
if(hex_to_long(b->tbox, 4) == BITS_PER_COMPONENT_BOX) {
println(INFO, "Bits Per Component box");
@@ -312,6 +322,8 @@
}
int jp2_parse_boxes(FILE *fd, type_image *img) {
+ box *ft;
+ box *b;
box *sig = get_next_box(fd);
if(hex_to_long(sig->tbox, 4) != JP2_SIGNATURE_BOX)
@@ -323,7 +335,7 @@
//dispose_of(sig);
//free(sig);
- box *ft = get_next_box(fd);
+ ft = get_next_box(fd);
if(hex_to_long(ft->tbox, 4) != JP2_FILETYPE_BOX)
println(INFO, "JP2 filetype box should directly follow JP2 signature box");
@@ -331,7 +343,7 @@
if(h_filetype_box(ft, img))
return 1;
- box *b;
+
while( (b = get_next_box(fd)) != NULL ) {
if(hex_to_long(b->tbox, 4) == JP2_HEADER_BOX) {
println(INFO, "Header Box");
Index: gpu_jpeg2k/jp2_decoder.c
===================================================================
--- gpu_jpeg2k/jp2_decoder.c (revision 84)
+++ gpu_jpeg2k/jp2_decoder.c (working copy)
@@ -40,12 +40,16 @@
int main(int argc, char **argv)
{
- println_start(INFO);
-
int i;
+ type_image *img;
type_parameters *param = (type_parameters*)malloc(sizeof(type_parameters));
type_tile *tile;
+ FILE *fsrc;
+ long file_length;
+ println_start(INFO);
+
+
if (argc < 2)
{
printf("Usage: decoder <image-file-name>\n\7");
@@ -56,15 +60,15 @@
println_var(INFO, "%s %s", argv[0], argv[1]);
- type_image *img = (type_image *)malloc(sizeof(type_image));
+ img = (type_image *)malloc(sizeof(type_image));
- FILE *fsrc = fopen(argv[1], "rb");
+ fsrc = fopen(argv[1], "rb");
if (!fsrc) {
println_var(INFO, "[ERROR] failed to open %s for reading\n", argv[1]);
return 1;
}
fseek(fsrc, 0, SEEK_END);
- long file_length = ftell(fsrc);
+ file_length = ftell(fsrc);
fseek(fsrc, 0, SEEK_SET);
Index: gpu_jpeg2k/klt/CMakeLists.txt
===================================================================
--- gpu_jpeg2k/klt/CMakeLists.txt (revision 84)
+++ gpu_jpeg2k/klt/CMakeLists.txt (working copy)
@@ -17,7 +17,6 @@
)
target_link_libraries(klt
- m
)
cuda_add_cublas_to_target(klt)
Index: gpu_jpeg2k/klt/adjust.cu
===================================================================
--- gpu_jpeg2k/klt/adjust.cu (revision 84)
+++ gpu_jpeg2k/klt/adjust.cu (working copy)
@@ -91,7 +91,7 @@
}
void test_cublas() {
- int size = 2;
+ const int size = 2;
int i = 0;
// If 'y' op(A)=A^T, else op(A)=A
char transa = 't';
Index: gpu_jpeg2k/preprocessing/preprocess_gpu.cu
===================================================================
--- gpu_jpeg2k/preprocessing/preprocess_gpu.cu (revision 84)
+++ gpu_jpeg2k/preprocessing/preprocess_gpu.cu (working copy)
@@ -46,6 +46,7 @@
#include "preprocessing_constants.cuh"
+
/**
* @brief CUDA kernel for DC level shifting coder.
*
@@ -300,6 +301,14 @@
void __global__ ict_kernel(type_data *img_r, type_data *img_g, type_data *img_b, const uint16_t width, const uint16_t height, const int level_shift) {
int i = threadIdx.x;
int j = threadIdx.y;
+ const float Wr = 0.299f;
+ const float Wb = 0.114f;
+ //const float Wg = 1 - Wr - Wb;
+ const float Wg = 1.0f - 0.299f - 0.114f;
+ const float Umax = 0.436f;
+ const float Vmax = 0.615f;
+
+
int n = i + blockIdx.x * TILE_SIZEX;
int m = j + blockIdx.y * TILE_SIZEY;
int idx = n + m * width;
@@ -373,6 +382,12 @@
void __global__ tci_kernel(type_data *img_r, type_data *img_g, type_data *img_b, const uint16_t width, const uint16_t height, const int level_shift, const int min, const int max) {
int i = threadIdx.x;
int j = threadIdx.y;
+ const float Wr = 0.299f;
+ const float Wb = 0.114f;
+ //const float Wg = 1 - Wr - Wb;
+ const float Wg = 1.0f - 0.299f - 0.114f;
+ const float Umax = 0.436f;
+ const float Vmax = 0.615f;
int n = i + blockIdx.x * TILE_SIZEX;
int m = j + blockIdx.y * TILE_SIZEY;
int idx = n + m * width;
Index: gpu_jpeg2k/print_info/print_info.c
===================================================================
--- gpu_jpeg2k/print_info/print_info.c (revision 84)
+++ gpu_jpeg2k/print_info/print_info.c (working copy)
@@ -1,115 +1,122 @@
-/*
-Copyright 2009-2013 Poznan Supercomputing and Networking Center
-
-Authors:
-Milosz Ciznicki miloszc@man.poznan.pl
-
-GPU JPEG2K is free software: you can redistribute it and/or modify
-it under the terms of the GNU Affero General Public License as published by
-the Free Software Foundation, either version 3 of the License, or
-(at your option) any later version.
-
-GPU JPEG2K is distributed in the hope that it will be useful,
-but WITHOUT ANY WARRANTY; without even the implied warranty of
-MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the
-GNU Affero General Public License for more details.
-
-You should have received a copy of the GNU Affero General Public License
-along with GPU JPEG2K. If not, see <http://www.gnu.org/licenses/>.
-*/
-/*
- * print_info.c
- *
- * Created on: 21-11-2010
- * Author: Milosz Ciznicki
- */
-#include <stdio.h>
-#include <stdarg.h>
-#include <sys/time.h>
-#include <time.h>
-
-struct timeval start_time;
-
-void println(const char* file, const char* function, const int line, const char *str)
-{
- struct timeval tv;
- struct timezone tz;
- struct tm *tm;
- gettimeofday(&tv, &tz);
- tm = localtime(&tv.tv_sec);
- fprintf(stdout, "%d:%02d:%02d %6ld [%s] (%s:%d) %s\n", tm->tm_hour, tm->tm_min, tm->tm_sec, tv.tv_usec, function, file, line, str);
-}
-
-void println_var(const char* file, const char* function, const int line, const char* format, ...)
-{
- char status[512];
- va_list arglist;
-
- va_start(arglist, format);
- vsprintf(status, format, arglist);
- va_end(arglist);
-
- println(file, function, line, status);
-}
-
-void println_start(const char* file, const char* function, const int line)
-{
- println(file, function, line, "start");
-}
-
-void println_end(const char* file, const char* function, const int line)
-{
- println(file, function, line, "end");
-}
-
-/*void start_measure()
-{
- gettimeofday(&start_time, NULL);
-}*/
-
-long int start_measure()
-{
- struct timeval start;
- gettimeofday(&start, NULL);
-
- return start.tv_sec * 1000000 + start.tv_usec;
-}
-
-long int stop_measure(long int start)
-{
- struct timeval end;
- gettimeofday(&end, NULL);
- long int time = (end.tv_sec * 1000000 + end.tv_usec) - start;
-
- return time;
-}
-
-/*long int stop_measure(const char* file, const char* function, const int line)
-{
- struct timeval end_time;
- gettimeofday(&end_time, NULL);
- long int time = (end_time.tv_sec - start_time.tv_sec) * 1000000 + end_time.tv_usec - start_time.tv_usec;
- println_var(file, function, line, "Computation time:%ld", time);
-
- return time;
-}*/
-
-long int stop_measure_msg(const char* file, const char* function, const int line, char *msg)
-{
- struct timeval end_time;
- gettimeofday(&end_time, NULL);
- long int time = (end_time.tv_sec - start_time.tv_sec) * 1000000 + end_time.tv_usec - start_time.tv_usec;
- println_var(file, function, line, "%s:%ld", msg, time);
-
- return time;
-}
-
-long int stop_measure_no_info()
-{
- struct timeval end_time;
- gettimeofday(&end_time, NULL);
- long int time = (end_time.tv_sec - start_time.tv_sec) * 1000000 + end_time.tv_usec - start_time.tv_usec;
- printf("%ld\n", time);
-
- return time;
-}
+/*
+Copyright 2009-2013 Poznan Supercomputing and Networking Center
+
+Authors:
+Milosz Ciznicki miloszc@man.poznan.pl
+
+GPU JPEG2K is free software: you can redistribute it and/or modify
+it under the terms of the GNU Affero General Public License as published by
+the Free Software Foundation, either version 3 of the License, or
+(at your option) any later version.
+
+GPU JPEG2K is distributed in the hope that it will be useful,
+but WITHOUT ANY WARRANTY; without even the implied warranty of
+MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the
+GNU Affero General Public License for more details.
+
+You should have received a copy of the GNU Affero General Public License
+along with GPU JPEG2K. If not, see <http://www.gnu.org/licenses/>.
+*/
+/*
+ * print_info.c
+ *
+ * Created on: 21-11-2010
+ * Author: Milosz Ciznicki
+ */
+#include <stdio.h>
+#include <stdarg.h>
+#if defined(_WIN32)
+ #include "timeval.h"
+#else
+ #include <sys/time.h>
+#endif
+#include <time.h>
+
+struct timeval start_time;
+
+void println(const char* file, const char* function, const int line, const char *str)
+{
+ struct timeval tv;
+ struct timezone tz;
+ struct tm *tm;
+ gettimeofday(&tv, &tz);
+ tm = localtime(&tv.tv_sec);
+ fprintf(stdout, "%d:%02d:%02d %6ld [%s] (%s:%d) %s\n", tm->tm_hour, tm->tm_min, tm->tm_sec, tv.tv_usec, function, file, line, str);
+}
+
+void println_var(const char* file, const char* function, const int line, const char* format, ...)
+{
+ char status[512];
+ va_list arglist;
+
+ va_start(arglist, format);
+ vsprintf(status, format, arglist);
+ va_end(arglist);
+
+ println(file, function, line, status);
+}
+
+void println_start(const char* file, const char* function, const int line)
+{
+ println(file, function, line, "start");
+}
+
+void println_end(const char* file, const char* function, const int line)
+{
+ println(file, function, line, "end");
+}
+
+/*void start_measure()
+{
+ gettimeofday(&start_time, NULL);
+}*/
+
+long int start_measure()
+{
+ struct timeval start;
+ gettimeofday(&start, NULL);
+
+ return start.tv_sec * 1000000 + start.tv_usec;
+}
+
+long int stop_measure(long int start)
+{
+ struct timeval end;
+ long int time;
+ gettimeofday(&end, NULL);
+ time = (end.tv_sec * 1000000 + end.tv_usec) - start;
+
+ return time;
+}
+
+/*long int stop_measure(const char* file, const char* function, const int line)
+{
+ struct timeval end_time;
+ gettimeofday(&end_time, NULL);
+ long int time = (end_time.tv_sec - start_time.tv_sec) * 1000000 + end_time.tv_usec - start_time.tv_usec;
+ println_var(file, function, line, "Computation time:%ld", time);
+
+ return time;
+}*/
+
+long int stop_measure_msg(const char* file, const char* function, const int line, char *msg)
+{
+ struct timeval end_time;
+ long int time;
+ gettimeofday(&end_time, NULL);
+ time = (end_time.tv_sec - start_time.tv_sec) * 1000000 + end_time.tv_usec - start_time.tv_usec;
+ println_var(file, function, line, "%s:%ld", msg, time);
+
+ return time;
+}
+
+long int stop_measure_no_info()
+{
+ struct timeval end_time;
+ long int time;
+ gettimeofday(&end_time, NULL);
+ time = (end_time.tv_sec - start_time.tv_sec) * 1000000 + end_time.tv_usec - start_time.tv_usec;
+ printf("%ld\n", time);
+
+ return time;
+}
Index: gpu_jpeg2k/tier1/coeff_coder/CMakeLists.txt
===================================================================
--- gpu_jpeg2k/tier1/coeff_coder/CMakeLists.txt (revision 84)
+++ gpu_jpeg2k/tier1/coeff_coder/CMakeLists.txt (working copy)
@@ -1,7 +1,7 @@
set(CUDA_NVCC_FLAGS ${CUDA_NVCC_FLAGS})
#"--ptxas-options=-dlcm=cg"
-find_package(PkgConfig REQUIRED)
+#find_package(PkgConfig REQUIRED)
#find_package(Boost REQUIRED)
#PKG_CHECK_MODULES(XML libxml++-2.6 REQUIRED)
Index: gpu_jpeg2k/tier1/coeff_coder/gpu_c2luts.cuh
===================================================================
--- gpu_jpeg2k/tier1/coeff_coder/gpu_c2luts.cuh (revision 84)
+++ gpu_jpeg2k/tier1/coeff_coder/gpu_c2luts.cuh (working copy)
@@ -1,22 +1,22 @@
-/*
-Copyright 2009-2013 Poznan Supercomputing and Networking Center
-
-Authors:
-Milosz Ciznicki miloszc@man.poznan.pl
-
-GPU JPEG2K is free software: you can redistribute it and/or modify
-it under the terms of the GNU Affero General Public License as published by
-the Free Software Foundation, either version 3 of the License, or
-(at your option) any later version.
-
-GPU JPEG2K is distributed in the hope that it will be useful,
-but WITHOUT ANY WARRANTY; without even the implied warranty of
-MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the
-GNU Affero General Public License for more details.
-
-You should have received a copy of the GNU Affero General Public License
-along with GPU JPEG2K. If not, see <http://www.gnu.org/licenses/>.
-*/
+/*
+Copyright 2009-2013 Poznan Supercomputing and Networking Center
+
+Authors:
+Milosz Ciznicki miloszc@man.poznan.pl
+
+GPU JPEG2K is free software: you can redistribute it and/or modify
+it under the terms of the GNU Affero General Public License as published by
+the Free Software Foundation, either version 3 of the License, or
+(at your option) any later version.
+
+GPU JPEG2K is distributed in the hope that it will be useful,
+but WITHOUT ANY WARRANTY; without even the implied warranty of
+MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the
+GNU Affero General Public License for more details.
+
+You should have received a copy of the GNU Affero General Public License
+along with GPU JPEG2K. If not, see <http://www.gnu.org/licenses/>.
+*/
#pragma once
__device__ __constant__ const unsigned short Qe[] = {
Index: gpu_jpeg2k/tier1/coeff_coder/gpu_mq-coder.cuh
===================================================================
--- gpu_jpeg2k/tier1/coeff_coder/gpu_mq-coder.cuh (revision 84)
+++ gpu_jpeg2k/tier1/coeff_coder/gpu_mq-coder.cuh (working copy)
@@ -1,28 +1,34 @@
-/*
-Copyright 2009-2013 Poznan Supercomputing and Networking Center
-
-Authors:
-Milosz Ciznicki miloszc@man.poznan.pl
-
-GPU JPEG2K is free software: you can redistribute it and/or modify
-it under the terms of the GNU Affero General Public License as published by
-the Free Software Foundation, either version 3 of the License, or
-(at your option) any later version.
-
-GPU JPEG2K is distributed in the hope that it will be useful,
-but WITHOUT ANY WARRANTY; without even the implied warranty of
-MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the
-GNU Affero General Public License for more details.
-
-You should have received a copy of the GNU Affero General Public License
-along with GPU JPEG2K. If not, see <http://www.gnu.org/licenses/>.
-*/
+/*
+Copyright 2009-2013 Poznan Supercomputing and Networking Center
+
+Authors:
+Milosz Ciznicki miloszc@man.poznan.pl
+
+GPU JPEG2K is free software: you can redistribute it and/or modify
+it under the terms of the GNU Affero General Public License as published by
+the Free Software Foundation, either version 3 of the License, or
+(at your option) any later version.
+
+GPU JPEG2K is distributed in the hope that it will be useful,
+but WITHOUT ANY WARRANTY; without even the implied warranty of
+MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the
+GNU Affero General Public License for more details.
+
+You should have received a copy of the GNU Affero General Public License
+along with GPU JPEG2K. If not, see <http://www.gnu.org/licenses/>.
+*/
#ifndef GPU_MQ_CODER_CUH_
#define GPU_MQ_CODER_CUH_
+
+#include <inttypes.h>
+#include <stdint.h>
#include "gpu_c2luts.cuh"
typedef unsigned char byte;
+#ifndef _WIN32
+typedef long long int __int64;
+#endif
#define CX_RUN 18
#define CX_UNI 17
@@ -298,8 +304,6 @@
//#include "gpu_emu_64_arit.cuh"
-typedef long long int __int64;
-
__device__ int calcFmin(MQEncoder &encoder)
{
__int64 sop = encoder.C;
Index: gpu_jpeg2k/tier1/ebcot/CMakeLists.txt
===================================================================
--- gpu_jpeg2k/tier1/ebcot/CMakeLists.txt (revision 84)
+++ gpu_jpeg2k/tier1/ebcot/CMakeLists.txt (working copy)
@@ -1,6 +1,6 @@
set(CUDA_NVCC_FLAGS ${CUDA_NVCC_FLAGS})
-find_package(PkgConfig REQUIRED)
+#find_package(PkgConfig REQUIRED)
cuda_add_library(ebcot
gpu_coeff_coder.cpp
Index: gpu_jpeg2k/tier1/quantizer.cu
===================================================================
--- gpu_jpeg2k/tier1/quantizer.cu (revision 84)
+++ gpu_jpeg2k/tier1/quantizer.cu (working copy)
@@ -145,7 +145,7 @@
if(img->wavelet_type)
{
/* The number of magnitude bits in the integer representation of the quantized data */
- sb->mag_bits = guard_bits - 1 + res_lvl->dec_lvl_no - (int)(log(base_step) / log(2));
+ sb->mag_bits = guard_bits - 1 + res_lvl->dec_lvl_no - (int)(log((double)base_step) / log((double)2));
/* Relative quantization step size. Step size is signaled relative to the wavelet coefficient bit depth. */
relative_step_size = base_step / (1 << res_lvl->dec_lvl_no);
shift_bits = 31 - sb->mag_bits;
Index: gpu_jpeg2k/tier2/CMakeLists.txt
===================================================================
--- gpu_jpeg2k/tier2/CMakeLists.txt (revision 84)
+++ gpu_jpeg2k/tier2/CMakeLists.txt (working copy)
@@ -9,7 +9,6 @@
)
TARGET_LINK_LIBRARIES(tier2
-m
misc
types
tier1
Index: gpu_jpeg2k/tier2/buffer.c
===================================================================
--- gpu_jpeg2k/tier2/buffer.c (revision 84)
+++ gpu_jpeg2k/tier2/buffer.c (working copy)
@@ -27,8 +27,9 @@
#include "buffer.h"
void init_dec_buffer(FILE *fsrc, type_buffer *src_buff) {
+ long file_length;
fseek(fsrc, 0, SEEK_END);
- long file_length = ftell(fsrc);
+ file_length = ftell(fsrc);
fseek(fsrc, 0, SEEK_SET);
src_buff->data = (uint8_t *) malloc(file_length);
Index: gpu_jpeg2k/tier2/codestream.c
===================================================================
--- gpu_jpeg2k/tier2/codestream.c (revision 84)
+++ gpu_jpeg2k/tier2/codestream.c (working copy)
@@ -272,13 +272,14 @@
void write_qcd_marker(type_buffer * buffer, type_image *img)
{
int i, j;
+ int lenght;
+ int qstyle;
write_short(buffer, QCD);
/* Lqcd - 4 + 3 * num_dlvls - no quantization
* - 5 - derived quantization
* - 5 + 6 * num_dlvls - expounded quantization */
/* XXX: Currently we use only derived quantization */
- int lenght;
- int qstyle;
+
if (img->wavelet_type == DWT_53) {
lenght = 4 + 3 * img->num_dlvls;
qstyle = 0;
@@ -336,6 +337,7 @@
{
for(j = 0; j < img->num_components; j++)
{
+ int expn, mant;
tile_comp = &(img->tile[i].tile_comp[j]);
seek_buffer(buffer, pos);
/* Sqcd - quantization style for all components*/
@@ -347,8 +349,6 @@
println_var(INFO, "Error: Unsupported quantization style: %d!", tile_comp->quant_style);
}
- int expn, mant;
-
if(tile_comp->quant_style == 0) /* No quantization */
{
for (k = 0; k < tile_comp->num_rlvls; k++) {
@@ -365,11 +365,12 @@
}
} else /* Derived quantization */
{
+ type_subband *sb_ll;
tmp = read_buffer(buffer, 2);
expn = tmp >> 11;
mant = tmp & 0x7ff;
- type_subband *sb_ll = &(tile_comp->res_lvls[0].subbands[0]);
+ sb_ll = &(tile_comp->res_lvls[0].subbands[0]);
sb_ll->expn = expn;
sb_ll->mant = mant;
@@ -726,8 +727,9 @@
tag_tree_reset(sb->inc_tt);
tag_tree_reset(sb->zero_bit_plane_tt);
for (j = 0; j < sb->num_cblks; j++) {
+ int tmp_zbp;
cblk = &(sb->cblks[j]);
- int tmp_zbp;
+
if (cblk->length == 0) {
tmp_zbp = sb->mag_bits;
} else {
Index: gpu_jpeg2k/tier2/codestream_mct.c
===================================================================
--- gpu_jpeg2k/tier2/codestream_mct.c (revision 84)
+++ gpu_jpeg2k/tier2/codestream_mct.c (working copy)
@@ -70,6 +70,8 @@
uint16_t Smct;
uint8_t type;
int i;
+ type_mct* old_mcts;
+ type_mct* mct;
/* Read MCT Marker */
marker = read_buffer(buffer, 2);
@@ -85,9 +87,9 @@
type = (Smct&(3<<4))>>4;
- type_mct* old_mcts = img->mct_data->mcts[type];
+ old_mcts = img->mct_data->mcts[type];
img->mct_data->mcts[type] = (type_mct*)realloc(img->mct_data->mcts[type], sizeof(type_mct) * (++img->mct_data->mcts_count[type]));
- type_mct* mct = &img->mct_data->mcts[type][img->mct_data->mcts_count[type]-1];
+ mct = &img->mct_data->mcts[type][img->mct_data->mcts_count[type]-1];
if(img->mct_data->mcts[type] == NULL) {
img->mct_data->mcts[type] = old_mcts;
@@ -117,7 +119,8 @@
void write_mcc_marker(type_buffer *buffer, type_mcc *mcc) {
int length;
int i,j;
-
+ type_mcc_data* data;
+
write_short(buffer, MCC);
length=5 + 6 * mcc->count;
@@ -128,7 +131,7 @@
write_int(buffer, length);
write_byte(buffer, mcc->index);
- type_mcc_data* data;
+
for(i=0; i<mcc->count; ++i) {
data = &mcc->data[i];
@@ -162,13 +165,14 @@
int count=0;
uint16_t temp_16;
uint8_t temp_8;
-
+ type_mcc* mcc;
+ type_mcc_data* mcc_data = NULL;
type_mcc* old_mccs = img->mct_data->mccs;
img->mct_data->mccs = (type_mcc*)realloc(img->mct_data->mccs, sizeof(type_mcc) * (++img->mct_data->mccs_count));
- type_mcc* mcc = &img->mct_data->mccs[img->mct_data->mccs_count-1];
+ mcc = &img->mct_data->mccs[img->mct_data->mccs_count-1];
mcc->data = NULL;
- type_mcc_data* mcc_data = NULL;
+
if(img->mct_data->mccs == NULL) {
img->mct_data->mccs = old_mccs;
@@ -250,7 +254,8 @@
void write_mic_marker(type_buffer *buffer, type_mic *mic) {
int length;
int i,j;
-
+ type_mic_data* data;
+
write_short(buffer, MIC);
length=5 + 5 * mic->count;
@@ -261,8 +266,6 @@
write_int(buffer, length);
write_byte(buffer, mic->index);
- type_mic_data* data;
-
for(i=0; i<mic->count; ++i) {
data = &mic->data[i];
write_short(buffer, data->input_count | ( data->input_component_type << 14));
@@ -290,10 +293,12 @@
int count=0;
uint16_t temp_16;
uint8_t temp_8;
+ type_mic* mic;
+ type_mic_data* mic_data;
type_mic* old_mics = img->mct_data->mics;
img->mct_data->mics = (type_mic*)realloc(img->mct_data->mics, sizeof(type_mic) * (++img->mct_data->mics_count));
- type_mic* mic = &img->mct_data->mics[img->mct_data->mics_count-1];
+ mic = &img->mct_data->mics[img->mct_data->mics_count-1];
if(img->mct_data->mics == NULL) {
img->mct_data->mics = old_mics;
@@ -302,7 +307,6 @@
return;
}
- type_mic_data* mic_data;
/* Read MCC Marker */
marker = read_buffer(buffer, 2);
@@ -366,7 +370,8 @@
int i;
int count;
uint16_t Satk;
-
+ uint8_t* data;
+
write_short(buffer, ATK);
if(atk->wavelet_type == 1) {
length = 8 + atk->coeff_type * ( atk->lifing_steps * atk->lifting_coefficients_per_step + atk->lifing_steps);
@@ -386,7 +391,7 @@
write_byte(buffer, atk->lifting_coefficients_per_step);
write_byte(buffer, atk->lifting_offset);
- uint8_t* data = atk->wavelet_type==0?atk->scaling_factor:atk->scaling_exponent;
+ data = atk->wavelet_type==0?atk->scaling_factor:atk->scaling_exponent;
for(i=0; i<(1<<atk->coeff_type); ++i) {
write_byte(buffer, data[i]);
}
@@ -416,9 +421,11 @@
int i;
int count;
uint16_t temp;
+ type_atk* atk;
+ uint8_t* t;
type_atk* old_atks = img->mct_data->atks;
img->mct_data->atks = (type_atk*)realloc(img->mct_data->atks, sizeof(type_atk) * (++img->mct_data->atk_count));
- type_atk* atk = &img->mct_data->atks[img->mct_data->atk_count-1];
+ atk = &img->mct_data->atks[img->mct_data->atk_count-1];
if(img->mct_data->atks == NULL) {
img->mct_data->atks = old_atks;
@@ -448,7 +455,7 @@
atk->lifting_offset = read_byte(buffer);
count = 1<<atk->coeff_type;
- uint8_t* t = (uint8_t*)calloc(count, sizeof(uint8_t));
+ t = (uint8_t*)calloc(count, sizeof(uint8_t));
for(i=0; i<count; ++i) {
t[i]=read_byte(buffer);
}
@@ -532,10 +539,10 @@
int length;
int i;
uint8_t temp;
-
+ type_ads* ads;
type_ads* old_adses = img->mct_data->adses;
img->mct_data->adses = (type_ads*)realloc(img->mct_data->adses, sizeof(type_ads) * (++img->mct_data->ads_count));
- type_ads* ads = &img->mct_data->adses[img->mct_data->ads_count-1];
+ ads = &img->mct_data->adses[img->mct_data->ads_count-1];
if(img->mct_data->adses == NULL) {
img->mct_data->adses = old_adses;
Index: gpu_jpeg2k/tier2/write_codestream.c
===================================================================
--- gpu_jpeg2k/tier2/write_codestream.c (revision 84)
+++ gpu_jpeg2k/tier2/write_codestream.c (working copy)
@@ -29,12 +29,13 @@
#include "write_codestream.h"
void write_codestream(type_image *img) {
+ FILE *fp;
type_buffer *buffer = (type_buffer *) malloc(sizeof(type_buffer));
init_buffer(buffer);
encode_codestream(buffer, img);
- FILE *fp = fopen(img->out_file, "wb");
+ fp = fopen(img->out_file, "wb");
write_buffer_to_file(buffer, fp);
Index: gpu_jpeg2k/types/buffered_stream.c
===================================================================
--- gpu_jpeg2k/types/buffered_stream.c (revision 84)
+++ gpu_jpeg2k/types/buffered_stream.c (working copy)
@@ -89,12 +89,13 @@
* @param buffer
*/
uint16_t peek_marker(type_buffer *buffer) {
+ uint16_t val;
if(buffer->bp+1 >= buffer->end)
{
println_var(INFO, "Error: Exceeded buffer bounds!");
exit(0);
}
- uint16_t val = (*buffer->bp)<<8;
+ val = (*buffer->bp)<<8;
val += *(buffer->bp+1);
return val;
}
Index: gpu_jpeg2k/types/image.c
===================================================================
--- gpu_jpeg2k/types/image.c (revision 84)
+++ gpu_jpeg2k/types/image.c (working copy)
@@ -26,7 +26,11 @@
*/
#include <stdio.h>
-#include <sys/time.h>
+#if defined(_WIN32)
+ #include <time.h>
+#else
+ #include <sys/time.h>
+#endif
#include <stdlib.h>
#include <errno.h>
Index: gpu_jpeg2k/types/image_bil.c
===================================================================
--- gpu_jpeg2k/types/image_bil.c (revision 84)
+++ gpu_jpeg2k/types/image_bil.c (working copy)
@@ -60,7 +60,7 @@
*/
static float *load_image(const char *image_filename, type_image_hyper *image_bil)
{
- println_start(INFO);
+
FILE *fp;
short *type_short_int;
double *type_double;
@@ -68,6 +68,8 @@
float *h_imagen;
int i;
+ println_start(INFO);
+
h_imagen = (float*) malloc(image_bil->num_lines * image_bil->num_samples * image_bil->num_bands * sizeof(float));
if (strstr(image_filename, ".bil") == NULL) {
@@ -95,10 +97,11 @@
}
break;
case 2: {
+ int read_lines;
type_short_int = (short *) malloc(image_bil->num_lines * image_bil->num_samples * image_bil->num_bands
* sizeof(short));
- int read_lines = fread(type_short_int, sizeof(short), (image_bil->num_lines * image_bil->num_samples * image_bil->num_bands), fp);
+ read_lines = fread(type_short_int, sizeof(short), (image_bil->num_lines * image_bil->num_samples * image_bil->num_bands), fp);
printf("read:%d orginal:%d\n", read_lines, image_bil->num_lines * image_bil->num_samples * image_bil->num_bands);
Index: gpu_jpeg2k/types/image_bip.c
===================================================================
--- gpu_jpeg2k/types/image_bip.c (revision 84)
+++ gpu_jpeg2k/types/image_bip.c (working copy)
@@ -60,7 +60,6 @@
*/
static float *load_image(const char *image_filename, type_image_hyper *image_bip)
{
- println_start(INFO);
FILE *fp;
short *type_short_int;
double *type_double;
@@ -68,6 +67,8 @@
float *h_imagen;
int i;
+ println_start(INFO);
+
h_imagen = (float*) malloc(image_bip->num_lines * image_bip->num_samples * image_bip->num_bands * sizeof(float));
if (strstr(image_filename, ".bip") == NULL) {
@@ -95,10 +96,11 @@
}
break;
case 2: {
+ int read_lines;
type_short_int = (short *) malloc(image_bip->num_lines * image_bip->num_samples * image_bip->num_bands
* sizeof(short));
- int read_lines = fread(type_short_int, sizeof(short), (image_bip->num_lines * image_bip->num_samples * image_bip->num_bands), fp);
+ read_lines = fread(type_short_int, sizeof(short), (image_bip->num_lines * image_bip->num_samples * image_bip->num_bands), fp);
printf("read:%d orginal:%d\n", read_lines, image_bip->num_lines * image_bip->num_samples * image_bip->num_bands);
Index: gpu_jpeg2k/types/image_bsq.c
===================================================================
--- gpu_jpeg2k/types/image_bsq.c (revision 84)
+++ gpu_jpeg2k/types/image_bsq.c (working copy)
@@ -68,6 +68,8 @@
float *type_float;
float *h_imagen;
int i;
+ int read_lines;
+ int spec;
h_imagen = (float*) malloc(image_bsq->num_lines * image_bsq->num_samples * image_bsq->num_bands * sizeof(float));
@@ -121,11 +123,11 @@
type_short_int = (short *) malloc(image_bsq->num_lines * image_bsq->num_samples * image_bsq->num_bands
* sizeof(short));
- int read_lines = fread(type_short_int, sizeof(short), (image_bsq->num_lines * image_bsq->num_samples * image_bsq->num_bands), fp);
+ read_lines = fread(type_short_int, sizeof(short), (image_bsq->num_lines * image_bsq->num_samples * image_bsq->num_bands), fp);
// printf("read:%d orginal:%d\n", read_lines, image_bsq->num_lines * image_bsq->num_samples * image_bsq->num_bands);
- int spec = 0;
+ spec = 0;
for (i = 0; i < image_bsq->num_lines * image_bsq->num_samples * image_bsq->num_bands; i++) {
h_imagen[i] = (float) type_short_int[i];
Index: gpu_jpeg2k/types/image_hyper.c
===================================================================
--- gpu_jpeg2k/types/image_hyper.c (revision 84)
+++ gpu_jpeg2k/types/image_hyper.c (working copy)
@@ -90,6 +90,10 @@
type_image *container = *_container;
type_image_hyper *image = (type_image_hyper*) malloc(sizeof(type_image_hyper));
type_data *data;
+ type_tile *tile;
+ type_tile_comp *tile_comp;
+ int band_size;
+ int x, y, c, i;
read_header(container->in_hfile, image);
@@ -106,15 +110,13 @@
//save_img_hyper(container, "cuprite_raw.raw");
- int band_size = container->width * container->height;
+ band_size = container->width * container->height;
// println_var(INFO, "Loaded %s: x:%d y:%d channels:%d depth:%d ", bsq_path, container->width, container->height,
// container->num_components, container->depth);
//copying data DIRECTLY as tiles to device
- int x, y, c, i;
- type_tile *tile;
- type_tile_comp *tile_comp;
+
for (i = 0; i < container->num_tiles; i++) {
tile = &(container->tile[i]);
for (c = 0; c < container->num_components; c++) {
@@ -140,6 +142,9 @@
int i, c, x, y;
type_tile *tile;
type_tile_comp *tile_comp;
+ int scan_width;
+ FILE *out_fp;
+
char** msg = (char**)calloc(1,sizeof(char*));
float *type_float;
@@ -147,11 +152,11 @@
type_float = (float *) malloc(img->width * img->height * img->num_components * sizeof(float));
if(type_float == NULL) {
- asprintf(msg, "Could not allocate data!");
+ //asprintf(msg, "Could not allocate data!");
perror(*msg);
}
- int scan_width = img->width;
+ scan_width = img->width;
// printf("num_tiles:%d num_comp:%d w:%d h:%d\n", img->num_tiles, img->num_components, img->width, img->height);
//exact opposite procedure as in read_img()
@@ -173,30 +178,30 @@
}
/* Read output file */
- FILE *out_fp = fopen(out_file, "wb");
+ out_fp = fopen(out_file, "wb");
printf("%d %d %d\n", img->width, img->height, img->num_components);
if(fwrite((void*)&(img->num_components), sizeof(unsigned short int), 1, out_fp)!=1) {
- asprintf(msg, "Read component");
+ //asprintf(msg, "Read component");
perror(*msg);
return;
}
if(fwrite((void*)&(img->width), sizeof(unsigned short int), 1, out_fp)!=1) {
- asprintf(msg, "Read component");
+ //asprintf(msg, "Read component");
perror(*msg);
return;
}
if(fwrite((void*)&(img->height), sizeof(unsigned short int), 1, out_fp)!=1) {
- asprintf(msg, "Read component");
+ //asprintf(msg, "Read component");
perror(*msg);
return;
}
if(fwrite(type_float, sizeof(float), img->width * img->height * img->num_components, out_fp)!=img->width * img->height * img->num_components) {
- asprintf(msg, "Read component");
+ //asprintf(msg, "Read component");
perror(*msg);
return;
}
@@ -222,6 +227,8 @@
void write_one_band(float *data, char *filename, type_image_hyper *image)
{
+ FILE *fp;
+
int size = image->num_lines * image->num_samples * sizeof(short int);
short int *odata = (short int*) malloc(size);
int i, j;
@@ -232,7 +239,6 @@
}
}
- FILE *fp;
if ((fp = fopen(filename, "wb")) == NULL) {
printf("ERROR %d. Can not write file: %s \n", errno, filename);
Index: gpu_jpeg2k/types/image_ordinary.c
===================================================================
--- gpu_jpeg2k/types/image_ordinary.c (revision 84)
+++ gpu_jpeg2k/types/image_ordinary.c (working copy)
@@ -1,504 +1,531 @@
-/*
-Copyright 2009-2013 Poznan Supercomputing and Networking Center
-
-Authors:
-Milosz Ciznicki miloszc@man.poznan.pl
-
-GPU JPEG2K is free software: you can redistribute it and/or modify
-it under the terms of the GNU Affero General Public License as published by
-the Free Software Foundation, either version 3 of the License, or
-(at your option) any later version.
-
-GPU JPEG2K is distributed in the hope that it will be useful,
-but WITHOUT ANY WARRANTY; without even the implied warranty of
-MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the
-GNU Affero General Public License for more details.
-
-You should have received a copy of the GNU Affero General Public License
-along with GPU JPEG2K. If not, see <http://www.gnu.org/licenses/>.
-*/
-/**
- * @file image.c
- * @brief This file include all the image loading/initialization/saving routines.
- *
- * @author Miłosz Ciżnicki
- * @author Jakub Misiorny <misiorny@man.poznan.pl>
- */
-
-#include <stdio.h>
-#include <sys/time.h>
-#include <stdlib.h>
-#include <errno.h>
-
-#include "../print_info/print_info.h"
-#include "image_ordinary.h"
-//#include "image_bsq.h"
-#include "../preprocessing/preprocess_gpu.h"
-#include "../misc/memory_management.cuh"
-#include "../tier2/markers.h"
-//#include "show_image.h"
-
-#include "../config/parameters.h"
-
-#include <FreeImage.h>
-//#include "dbg_image.h"
-
-//#define READ_TIME
-
-/**
- * @brief Initializes image container
- *
- * @param dib Bitmap opened by FreeImage
- * @param container malloc'ed type_image to be initialized.
- */
-void init_image(FIBITMAP* dib, type_image *container, type_parameters *param)
-{
- // println_start(INFO);
- container->height = FreeImage_GetHeight(dib);
- container->width = FreeImage_GetWidth(dib);
- container->depth = FreeImage_GetBPP(dib);
- container->num_components = get_num_comp(dib);
- container->num_range_bits = container->depth / container->num_components;
-
- /* TODO: check if origin data was signed? */
- container->sign = get_img_type(dib);
- container->num_dlvls = param->param_tile_comp_dlvls;
-
- set_coding_parameters(container, param);
- init_tiles(&container, param);
-
-// println_end(INFO);
-}
-
-uint8_t get_num_comp(FIBITMAP* dib)
-{
- FREE_IMAGE_COLOR_TYPE color_type = FreeImage_GetColorType(dib);
- switch(color_type)
- {
- case FIC_MINISWHITE: /*printf("FIC_MINISWHITE\n");*/ return 1U;
- case FIC_MINISBLACK: /*printf("FIC_MINISBLACK\n");*/ return 1U;
- case FIC_RGB: /*printf("RGB\n");*/ return 3U;
- default: printf("Unsupported image type!\n"); exit(0);
- /*case FIC_PALETTE: printf("FIC_PALETTE\n"); break;
- case FIC_RGBALPHA: printf("FIC_RGBALPHA\n"); break;
- case FIC_CMYK: printf("FIC_CMYK\n"); break;*/
- }
-}
-
-uint8_t get_img_type(FIBITMAP* dib)
-{
- FREE_IMAGE_TYPE image_type = FreeImage_GetImageType(dib);
-
- switch(image_type)
- {
- case FIT_BITMAP:
- case FIT_UINT16:
- case FIT_UINT32:
- case FIT_RGB16:
- case FIT_RGBA16: return UNSIGNED;
- case FIT_INT16:
- case FIT_INT32: return SIGNED;
- }
- return 2U;
-}
-
-/**
-FreeImage error handler
-@param fif Format / Plugin responsible for the error
-@param message Error message
-*/
-void FreeImageErrorHandler(FREE_IMAGE_FORMAT fif, const char *message) {
- printf("\n*** ");
- if(fif != FIF_UNKNOWN) {
- printf("%s Format\n", FreeImage_GetFormatFromFIF(fif));
- }
- printf("%s", message);
- printf(" ***\n");
-}
-
-/*void convert(const char *filename)
-{
- int ret;
- char cmdline[1048];
-
- char temp_file[32];
- tmpnam(temp_file); //get temp filename
-
- //printf("temp file is: '%s'\n", temp_file);
-
- //Create Windows bitmap 3.0 with 24bpp
- sprintf(cmdline, "%s \"%s\" -type truecolor \"BMP3:%s\"", "convert", filename, temp_file);
- //printf("cmdline = '%s'\n", cmdline);
- ret = system(cmdline);
- //system("pause");
- if(ret != 0) {
- printf("Could not convert '%s' to temp BMP '%s' !\n", filename, temp_file);
- printf("convert returned error %d\n", ret);
- printf("cmdline was: '%s'\n", cmdline);
- free(temp_file);
- exit(0);
- }
- strcpy(filename, temp_file);
-}*/
-
-
-/**
- * @brief Read image from file.
- *
- * @param path Path to image file.
- * @param _container Pointer to a malloc'ed memory for type_image.
- */
-long int read_ordinary_image(type_image **_container, type_parameters *param)
-{
- // println_start(INFO);
-#ifdef READ_TIME
- long int start_load;
- start_load = start_measure();
-#endif
-
- type_image *container = *_container;
- FREE_IMAGE_FORMAT formato = FreeImage_GetFileType(container->in_file, 0);
- FIBITMAP* dib = FreeImage_Load(formato, container->in_file, 0);
- FREE_IMAGE_TYPE image_type = FreeImage_GetImageType(dib);
- long int copy_time = 0;
-
-#ifdef READ_TIME
- cudaThreadSynchronize();
- printf("Load img:%ld\n", stop_measure(start_load));
-#endif
-
- FreeImage_SetOutputMessage(FreeImageErrorHandler);
-
- if(image_type == FIT_BITMAP)
- {
-// printf("BITMAP\n");
-#ifdef READ_TIME
- long int start_convert;
- start_convert = start_measure();
-#endif
-
- dib = FreeImage_ConvertTo24Bits(dib);
-
-#ifdef READ_TIME
- cudaThreadSynchronize();
- printf("Convert img:%ld\n", stop_measure(start_convert));
-#endif
-
- if(FreeImage_HasPixels(dib) == FALSE)
- {
- printf("Do not have pixel data!\n");
- exit(0);
- }
-
-#ifdef READ_TIME
- long int start_read;
- start_read = start_measure();
-#endif
-
- init_image(dib, container, param);
-
-#ifdef READ_TIME
- cudaThreadSynchronize();
- printf("Init img:%ld\n", stop_measure(start_read));
-#endif
-
-// println_var(INFO, "Loaded %s: x:%d y:%d channels:%d depth:%d ",
-// path, container->width, container->height, container->num_components, container->depth);
-
-#ifdef READ_TIME
- long int start_raw;
- start_raw = start_measure();
-#endif
-
- int scan_width = FreeImage_GetPitch(dib);
- int mem_size = container->height * scan_width;
-
- BYTE *bits = malloc(mem_size * sizeof(BYTE));
-
- // convert the bitmap to raw bits (top-left pixel first)
- FreeImage_ConvertToRawBits(bits, dib, scan_width, FreeImage_GetBPP(dib), FI_RGBA_RED_MASK/*FI_RGBA_BLUE_MASK*/, FI_RGBA_GREEN_MASK, /*FI_RGBA_RED_MASK*/FI_RGBA_BLUE_MASK, TRUE);
-// FreeImage_ConvertToRawBits(bits, dib, scan_width, FreeImage_GetBPP(dib), FI_RGBA_GREEN_MASK/*FI_RGBA_BLUE_MASK*/, FI_RGBA_BLUE_MASK, /*FI_RGBA_RED_MASK*/FI_RGBA_RED_MASK, TRUE);
-
- FreeImage_Unload(dib);
-#ifdef READ_TIME
- cudaThreadSynchronize();
- printf("Raw bits img:%ld\n", stop_measure(start_raw));
-#endif
-
- //FreeImage_Unload(dib);
-
- long int start_copy;
- start_copy = start_measure();
-
- //copying data DIRECTLY as tiles to device
- int x, y, c, i;
- type_tile *tile;
- type_tile_comp *tile_comp;
- for(i = 0; i < container->num_tiles; i++) {
- tile = &(container->tile[i]);
- for(c = 0; c < container->num_components; c++) {
- tile_comp = &(tile->tile_comp[c]);
- for(y = 0; y < tile_comp->height; y++) {
- for(x = 0; x < tile_comp->width; x++) {
- tile_comp->img_data[x + y*tile_comp->width] =
- (type_data)bits[(tile->tly + y) * scan_width + (tile->tlx + x) * container->num_components + c];
-// if(c == 0)
-// printf("%6d,", bits[(tile->tly + y) * scan_width + (tile->tlx + x) * container->num_components + c] - 128);
- }
-// if(c == 0)
-// printf("\n");
- }
- cuda_memcpy_htd(tile_comp->img_data, tile_comp->img_data_d, tile_comp->width * tile_comp->height * sizeof(type_data));
- cuda_h_free(tile_comp->img_data);
-// free(tile_comp->img_data);
- }
- }
-
-// cudaThreadSynchronize();
- copy_time = stop_measure(start_copy);
-
-/* char buff[128];
-
- sprintf(buff, "%s.raw\0", path);
-
- save_raw(container, buff);*/
-
-#ifdef READ_TIME
- printf("Copy img:%ld\n", copy_time);
-#endif
-
- //free(bits);
- } else if(image_type == FIT_RGB16 || image_type == FIT_UINT16)
- {
-// printf("RGB16 or UINT16\n");
- init_image(dib, container, param);
-
- int scan_width = FreeImage_GetPitch(dib)/sizeof(unsigned short);
-
- println_var(INFO, "Loaded %s: x:%d y:%d channels:%d depth:%d ",
- container->in_file, container->width, container->height, container->num_components, container->depth);
-
- //copying data DIRECTLY as tiles to device
- int x, y, c, i;
- type_tile *tile;
- type_tile_comp *tile_comp;
- for(i = 0; i < container->num_tiles; i++) {
- tile = &(container->tile[i]);
- for(c = 0; c < container->num_components; c++) {
- tile_comp = &(tile->tile_comp[c]);
- for(y = 0; y < tile_comp->height; y++) {
- for(x = 0; x < tile_comp->width; x++) {
- tile_comp->img_data[x + (tile_comp->height - 1 - y)*tile_comp->width] =
- (type_data)((unsigned short *)FreeImage_GetBits(dib))[(tile->tly + y) * scan_width + (tile->tlx + x) * container->num_components + c];
- }
- }
- cuda_memcpy_htd(tile_comp->img_data, tile_comp->img_data_d, tile_comp->width * tile_comp->height * sizeof(type_data));
- cuda_h_free(tile_comp->img_data);
- }
- }
- FreeImage_Unload(dib);
- } else
- {
- init_image(dib, container, param);
- dib = FreeImage_ConvertToType(dib, FIT_DOUBLE, TRUE);
-
- if(dib == NULL)
- {
- printf("Conversion not allowed!\n");
- exit(0);
- }
-
- int scan_width = FreeImage_GetPitch(dib)/sizeof(double);
-
- if(FreeImage_HasPixels(dib) == FALSE)
- {
- printf("Do not have pixel data!\n");
- exit(0);
- }
- println_var(INFO, "Loaded %s: x:%d y:%d channels:%d depth:%d ",
- container->in_file, container->width, container->height, container->num_components, container->depth);
-
- //copying data DIRECTLY as tiles to device
- int x, y, c, i;
- type_tile *tile;
- type_tile_comp *tile_comp;
- for(i = 0; i < container->num_tiles; i++) {
- tile = &(container->tile[i]);
- for(c = 0; c < container->num_components; c++) {
- tile_comp = &(tile->tile_comp[c]);
- for(y = 0; y < tile_comp->height; y++) {
- for(x = 0; x < tile_comp->width; x++) {
- tile_comp->img_data[x + (tile_comp->height - 1 - y)*tile_comp->width] =
- (type_data)((double *)FreeImage_GetBits(dib))[(tile->tly + y) * scan_width + (tile->tlx + x) * container->num_components + c];
- }
- }
- cuda_memcpy_htd(tile_comp->img_data, tile_comp->img_data_d, tile_comp->width * tile_comp->height * sizeof(type_data));
- cuda_h_free(tile_comp->img_data);
- }
- }
- FreeImage_Unload(dib);
- }
-
- return 0;
-}
-
-void save_tile_comp(type_tile_comp *tile_comp, char *filename)
-{
- float *image;
- int size = tile_comp->width * tile_comp->height;
- image = (float *) malloc(size * sizeof(float));
- cuda_memcpy_dth(tile_comp->img_data_d, image, size * sizeof(float));
-
- short int *odata = (short int*) malloc(size * sizeof(short int));
- int i, j;
-
- for (j = 0; j < tile_comp->height; j++) {
- for (i = 0; i < tile_comp->width; i++) {
- odata[i + j * tile_comp->width] = (short int) image[i + j * tile_comp->width];
- }
- }
-
- FILE *fp;
-
- if ((fp = fopen(filename, "wb")) == NULL) {
- printf("ERROR %d. Can not write file: %s \n", errno, filename);
- system("PAUSE");
- exit(1);
- } else {
- fseek(fp, 0L, SEEK_SET);
- fwrite(odata, sizeof(short int), size * sizeof(short int), fp);
- }
- fclose(fp);
- free(odata);
- free(image);
-}
-
-void save_tile_comp_with_shift(type_tile_comp *tile_comp, char *filename, int shift)
-{
- float *odata;
- int size = tile_comp->width * tile_comp->height * sizeof(float);
- odata = (float *) malloc(size);
- cuda_memcpy_dth(tile_comp->img_data_d, odata, size);
- int i, j;
-
- for (j = 0; j < tile_comp->height; j++) {
- for (i = 0; i < tile_comp->width; i++) {
- odata[i + j * tile_comp->width] += 1 << (shift - 1);
- printf("%.1f,", odata[i + j * tile_comp->width]);
- }
- printf("\n");
- }
-
- FILE *fp;
- if ((fp = fopen(filename, "wb")) == NULL) {
- printf("ERROR %d. Can not open file: %s \n", errno, filename);
- system("PAUSE");
- exit(1);
- } else {
- fseek(fp, 0L, SEEK_SET);
- fwrite(odata, sizeof(float), size, fp);
- }
- fclose(fp);
- free(odata);
-}
-
-int save_img_ord(type_image *img, const char *filename)
-{
- println_start(INFO);
- int i, c, x, y;
- type_tile *tile;
- type_tile_comp *tile_comp;
-
- int scan_width = img->width * (img->num_components);
- BYTE *bits = (BYTE*) malloc(img->height * scan_width * sizeof(BYTE));
-
- // printf("for\n");
- //exact opposite procedure as in read_img()
- for (i = 0; i < img->num_tiles; i++) {
- tile = &(img->tile[i]);
- for (c = 0; c < img->num_components; c++) {
- tile_comp = &(tile->tile_comp[c]);
- cuda_h_allocate_mem((void **) &(tile_comp->img_data), tile_comp->width * tile_comp->height
- * sizeof(type_data));
- cuda_memcpy_dth(tile_comp->img_data_d, tile_comp->img_data, tile_comp->width * tile_comp->height
- * sizeof(type_data));
- for (x = 0; x < tile_comp->width; x++) {
- for (y = 0; y < tile_comp->height; y++) {
- bits[(tile->tly + y) * scan_width + (tile->tlx + x) * img->num_components + c]
- = (BYTE) tile_comp->img_data[x + y * tile_comp->width];
-
- // if(tile_comp->img_data[x + y * tile_comp->width] > (BYTE) tile_comp->img_data[x + y * tile_comp->width]) {
- // printf("%f = %i\n", tile_comp->img_data[x + y * tile_comp->width], bits[(tile->tly + y) * scan_width + (tile->tlx + x) * img->num_components + c]);
- // }
- }
- }
- }
- }
-
- // convert the bitmap to raw bits (top-left pixel first)
- FIBITMAP *dst = FreeImage_ConvertFromRawBits(bits, img->width, img->height, scan_width, 24, FI_RGBA_RED_MASK,
- FI_RGBA_GREEN_MASK, FI_RGBA_BLUE_MASK, TRUE);
-
- if (FreeImage_Save(FIF_BMP, dst, filename, 0)) {
- println_var(INFO, "saved: %s", filename);
- } else {
- println_var(INFO, "saving FAILED: %s", filename);
- }
-
- FreeImage_Unload(dst);
- free(bits);
-
- return 0;
-
-}
-
-int save_img_grayscale(type_image *img, char *filename)
-{
- println_start(INFO);
- int i, c, x, y;
- type_tile *tile;
- type_tile_comp *tile_comp;
-
- tile = &(img->tile[0]);
- tile_comp = &(tile->tile_comp[0]);
- cuda_h_allocate_mem((void **) &(tile_comp->img_data), tile_comp->width * tile_comp->height
- * sizeof(type_data));
- cuda_memcpy_dth(tile_comp->img_data_d, tile_comp->img_data, tile_comp->width * tile_comp->height
- * sizeof(type_data));
-
- int scan_width = img->width * (img->num_components);
- BYTE *bits = (BYTE*) malloc(img->height * scan_width * sizeof(BYTE));
-
- // printf("for\n");
- //exact opposite procedure as in read_img()
- for (i = 0; i < img->num_tiles; i++) {
- for (c = 0; c < img->num_components; c++) {
- for (x = 0; x < tile_comp->width; x++) {
- for (y = 0; y < tile_comp->height; y++) {
- bits[(tile->tly + y) * scan_width + (tile->tlx + x) * img->num_components + c]
- = (BYTE) tile_comp->img_data[x + y * tile_comp->width];
-
- // if(tile_comp->img_data[x + y * tile_comp->width] > (BYTE) tile_comp->img_data[x + y * tile_comp->width]) {
- // printf("%f = %i\n", tile_comp->img_data[x + y * tile_comp->width], bits[(tile->tly + y) * scan_width + (tile->tlx + x) * img->num_components + c]);
- // }
- }
- }
- }
- }
-
- // convert the bitmap to raw bits (top-left pixel first)
- FIBITMAP *dst = FreeImage_ConvertFromRawBits(bits, img->width, img->height, scan_width, 24, FI_RGBA_RED_MASK,
- FI_RGBA_GREEN_MASK, FI_RGBA_BLUE_MASK, TRUE);
-
- if (FreeImage_Save(FIF_BMP, dst, filename, 0)) {
- println_var(INFO, "saved: %s", filename);
- } else {
- println_var(INFO, "saving FAILED: %s", filename);
- }
-
- FreeImage_Unload(dst);
- free(bits);
-
- return 0;
-
-}
+/*
+Copyright 2009-2013 Poznan Supercomputing and Networking Center
+
+Authors:
+Milosz Ciznicki miloszc@man.poznan.pl
+
+GPU JPEG2K is free software: you can redistribute it and/or modify
+it under the terms of the GNU Affero General Public License as published by
+the Free Software Foundation, either version 3 of the License, or
+(at your option) any later version.
+
+GPU JPEG2K is distributed in the hope that it will be useful,
+but WITHOUT ANY WARRANTY; without even the implied warranty of
+MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the
+GNU Affero General Public License for more details.
+
+You should have received a copy of the GNU Affero General Public License
+along with GPU JPEG2K. If not, see <http://www.gnu.org/licenses/>.
+*/
+/**
+ * @file image.c
+ * @brief This file include all the image loading/initialization/saving routines.
+ *
+ * @author Miłosz Ciżnicki
+ * @author Jakub Misiorny <misiorny@man.poznan.pl>
+ */
+
+#include <stdio.h>
+#if defined(_WIN32)
+ #include <time.h>
+#else
+ #include <sys/time.h>
+#endif
+#include <stdlib.h>
+#include <errno.h>
+
+#include "../print_info/print_info.h"
+#include "image_ordinary.h"
+//#include "image_bsq.h"
+#include "../preprocessing/preprocess_gpu.h"
+#include "../misc/memory_management.cuh"
+#include "../tier2/markers.h"
+//#include "show_image.h"
+
+#include "../config/parameters.h"
+
+#include <FreeImage.h>
+//#include "dbg_image.h"
+
+//#define READ_TIME
+
+/**
+ * @brief Initializes image container
+ *
+ * @param dib Bitmap opened by FreeImage
+ * @param container malloc'ed type_image to be initialized.
+ */
+void init_image(FIBITMAP* dib, type_image *container, type_parameters *param)
+{
+ // println_start(INFO);
+ container->height = FreeImage_GetHeight(dib);
+ container->width = FreeImage_GetWidth(dib);
+ container->depth = FreeImage_GetBPP(dib);
+ container->num_components = get_num_comp(dib);
+ container->num_range_bits = container->depth / container->num_components;
+
+ /* TODO: check if origin data was signed? */
+ container->sign = get_img_type(dib);
+ container->num_dlvls = param->param_tile_comp_dlvls;
+
+ set_coding_parameters(container, param);
+ init_tiles(&container, param);
+
+// println_end(INFO);
+}
+
+uint8_t get_num_comp(FIBITMAP* dib)
+{
+ FREE_IMAGE_COLOR_TYPE color_type = FreeImage_GetColorType(dib);
+ switch(color_type)
+ {
+ case FIC_MINISWHITE: /*printf("FIC_MINISWHITE\n");*/ return 1U;
+ case FIC_MINISBLACK: /*printf("FIC_MINISBLACK\n");*/ return 1U;
+ case FIC_RGB: /*printf("RGB\n");*/ return 3U;
+ default: printf("Unsupported image type!\n"); exit(0);
+ /*case FIC_PALETTE: printf("FIC_PALETTE\n"); break;
+ case FIC_RGBALPHA: printf("FIC_RGBALPHA\n"); break;
+ case FIC_CMYK: printf("FIC_CMYK\n"); break;*/
+ }
+}
+
+uint8_t get_img_type(FIBITMAP* dib)
+{
+ FREE_IMAGE_TYPE image_type = FreeImage_GetImageType(dib);
+
+ switch(image_type)
+ {
+ case FIT_BITMAP:
+ case FIT_UINT16:
+ case FIT_UINT32:
+ case FIT_RGB16:
+ case FIT_RGBA16: return UNSIGNED;
+ case FIT_INT16:
+ case FIT_INT32: return SIGNED;
+ }
+ return 2U;
+}
+
+/**
+FreeImage error handler
+@param fif Format / Plugin responsible for the error
+@param message Error message
+*/
+void FreeImageErrorHandler(FREE_IMAGE_FORMAT fif, const char *message) {
+ printf("\n*** ");
+ if(fif != FIF_UNKNOWN) {
+ printf("%s Format\n", FreeImage_GetFormatFromFIF(fif));
+ }
+ printf("%s", message);
+ printf(" ***\n");
+}
+
+/*void convert(const char *filename)
+{
+ int ret;
+ char cmdline[1048];
+
+ char temp_file[32];
+ tmpnam(temp_file); //get temp filename
+
+ //printf("temp file is: '%s'\n", temp_file);
+
+ //Create Windows bitmap 3.0 with 24bpp
+ sprintf(cmdline, "%s \"%s\" -type truecolor \"BMP3:%s\"", "convert", filename, temp_file);
+ //printf("cmdline = '%s'\n", cmdline);
+ ret = system(cmdline);
+ //system("pause");
+ if(ret != 0) {
+ printf("Could not convert '%s' to temp BMP '%s' !\n", filename, temp_file);
+ printf("convert returned error %d\n", ret);
+ printf("cmdline was: '%s'\n", cmdline);
+ free(temp_file);
+ exit(0);
+ }
+ strcpy(filename, temp_file);
+}*/
+
+
+/**
+ * @brief Read image from file.
+ *
+ * @param path Path to image file.
+ * @param _container Pointer to a malloc'ed memory for type_image.
+ */
+long int read_ordinary_image(type_image **_container, type_parameters *param)
+{
+ // println_start(INFO);
+ int scan_width;
+ int mem_size;
+ BYTE *bits;
+ long int start_copy;
+ int x, y, c, i;
+ type_tile *tile;
+ type_tile_comp *tile_comp;
+#ifdef READ_TIME
+ long int start_load;
+ start_load = start_measure();
+#endif
+
+ type_image *container = *_container;
+ FREE_IMAGE_FORMAT formato = FreeImage_GetFileType(container->in_file, 0);
+ FIBITMAP* dib = FreeImage_Load(formato, container->in_file, 0);
+ FREE_IMAGE_TYPE image_type = FreeImage_GetImageType(dib);
+ long int copy_time = 0;
+
+#ifdef READ_TIME
+ cudaThreadSynchronize();
+ printf("Load img:%ld\n", stop_measure(start_load));
+#endif
+
+ FreeImage_SetOutputMessage(FreeImageErrorHandler);
+
+ if(image_type == FIT_BITMAP)
+ {
+// printf("BITMAP\n");
+#ifdef READ_TIME
+ long int start_convert;
+ start_convert = start_measure();
+#endif
+
+ dib = FreeImage_ConvertTo24Bits(dib);
+
+#ifdef READ_TIME
+ cudaThreadSynchronize();
+ printf("Convert img:%ld\n", stop_measure(start_convert));
+#endif
+
+ if(FreeImage_HasPixels(dib) == FALSE)
+ {
+ printf("Do not have pixel data!\n");
+ exit(0);
+ }
+
+#ifdef READ_TIME
+ long int start_read;
+ start_read = start_measure();
+#endif
+
+ init_image(dib, container, param);
+
+#ifdef READ_TIME
+ cudaThreadSynchronize();
+ printf("Init img:%ld\n", stop_measure(start_read));
+#endif
+
+// println_var(INFO, "Loaded %s: x:%d y:%d channels:%d depth:%d ",
+// path, container->width, container->height, container->num_components, container->depth);
+
+#ifdef READ_TIME
+ long int start_raw;
+ start_raw = start_measure();
+#endif
+
+ scan_width = FreeImage_GetPitch(dib);
+ mem_size = container->height * scan_width;
+
+ bits = malloc(mem_size * sizeof(BYTE));
+
+ // convert the bitmap to raw bits (top-left pixel first)
+ FreeImage_ConvertToRawBits(bits, dib, scan_width, FreeImage_GetBPP(dib), FI_RGBA_RED_MASK/*FI_RGBA_BLUE_MASK*/, FI_RGBA_GREEN_MASK, /*FI_RGBA_RED_MASK*/FI_RGBA_BLUE_MASK, TRUE);
+// FreeImage_ConvertToRawBits(bits, dib, scan_width, FreeImage_GetBPP(dib), FI_RGBA_GREEN_MASK/*FI_RGBA_BLUE_MASK*/, FI_RGBA_BLUE_MASK, /*FI_RGBA_RED_MASK*/FI_RGBA_RED_MASK, TRUE);
+
+ FreeImage_Unload(dib);
+#ifdef READ_TIME
+ cudaThreadSynchronize();
+ printf("Raw bits img:%ld\n", stop_measure(start_raw));
+#endif
+
+ //FreeImage_Unload(dib);
+
+
+ start_copy = start_measure();
+
+ //copying data DIRECTLY as tiles to device
+
+
+ for(i = 0; i < container->num_tiles; i++) {
+ tile = &(container->tile[i]);
+ for(c = 0; c < container->num_components; c++) {
+ tile_comp = &(tile->tile_comp[c]);
+ for(y = 0; y < tile_comp->height; y++) {
+ for(x = 0; x < tile_comp->width; x++) {
+ tile_comp->img_data[x + y*tile_comp->width] =
+ (type_data)bits[(tile->tly + y) * scan_width + (tile->tlx + x) * container->num_components + c];
+// if(c == 0)
+// printf("%6d,", bits[(tile->tly + y) * scan_width + (tile->tlx + x) * container->num_components + c] - 128);
+ }
+// if(c == 0)
+// printf("\n");
+ }
+ cuda_memcpy_htd(tile_comp->img_data, tile_comp->img_data_d, tile_comp->width * tile_comp->height * sizeof(type_data));
+ cuda_h_free(tile_comp->img_data);
+// free(tile_comp->img_data);
+ }
+ }
+
+// cudaThreadSynchronize();
+ copy_time = stop_measure(start_copy);
+
+/* char buff[128];
+
+ sprintf(buff, "%s.raw\0", path);
+
+ save_raw(container, buff);*/
+
+#ifdef READ_TIME
+ printf("Copy img:%ld\n", copy_time);
+#endif
+
+ //free(bits);
+ } else if(image_type == FIT_RGB16 || image_type == FIT_UINT16)
+ {
+// printf("RGB16 or UINT16\n");
+ int scan_width;
+ int x, y, c, i;
+ type_tile *tile;
+ type_tile_comp *tile_comp;
+ init_image(dib, container, param);
+
+ scan_width = FreeImage_GetPitch(dib)/sizeof(unsigned short);
+
+ println_var(INFO, "Loaded %s: x:%d y:%d channels:%d depth:%d ",
+ container->in_file, container->width, container->height, container->num_components, container->depth);
+
+ //copying data DIRECTLY as tiles to device
+
+
+ for(i = 0; i < container->num_tiles; i++) {
+ tile = &(container->tile[i]);
+ for(c = 0; c < container->num_components; c++) {
+ tile_comp = &(tile->tile_comp[c]);
+ for(y = 0; y < tile_comp->height; y++) {
+ for(x = 0; x < tile_comp->width; x++) {
+ tile_comp->img_data[x + (tile_comp->height - 1 - y)*tile_comp->width] =
+ (type_data)((unsigned short *)FreeImage_GetBits(dib))[(tile->tly + y) * scan_width + (tile->tlx + x) * container->num_components + c];
+ }
+ }
+ cuda_memcpy_htd(tile_comp->img_data, tile_comp->img_data_d, tile_comp->width * tile_comp->height * sizeof(type_data));
+ cuda_h_free(tile_comp->img_data);
+ }
+ }
+ FreeImage_Unload(dib);
+ } else
+ {
+ int scan_width;
+ int x, y, c, i;
+ type_tile *tile;
+ type_tile_comp *tile_comp;
+
+ init_image(dib, container, param);
+ dib = FreeImage_ConvertToType(dib, FIT_DOUBLE, TRUE);
+
+ if(dib == NULL)
+ {
+ printf("Conversion not allowed!\n");
+ exit(0);
+ }
+
+ scan_width = FreeImage_GetPitch(dib)/sizeof(double);
+
+ if(FreeImage_HasPixels(dib) == FALSE)
+ {
+ printf("Do not have pixel data!\n");
+ exit(0);
+ }
+ println_var(INFO, "Loaded %s: x:%d y:%d channels:%d depth:%d ",
+ container->in_file, container->width, container->height, container->num_components, container->depth);
+
+ //copying data DIRECTLY as tiles to device
+
+ for(i = 0; i < container->num_tiles; i++) {
+ tile = &(container->tile[i]);
+ for(c = 0; c < container->num_components; c++) {
+ tile_comp = &(tile->tile_comp[c]);
+ for(y = 0; y < tile_comp->height; y++) {
+ for(x = 0; x < tile_comp->width; x++) {
+ tile_comp->img_data[x + (tile_comp->height - 1 - y)*tile_comp->width] =
+ (type_data)((double *)FreeImage_GetBits(dib))[(tile->tly + y) * scan_width + (tile->tlx + x) * container->num_components + c];
+ }
+ }
+ cuda_memcpy_htd(tile_comp->img_data, tile_comp->img_data_d, tile_comp->width * tile_comp->height * sizeof(type_data));
+ cuda_h_free(tile_comp->img_data);
+ }
+ }
+ FreeImage_Unload(dib);
+ }
+
+ return 0;
+}
+
+void save_tile_comp(type_tile_comp *tile_comp, char *filename)
+{
+ float *image;
+ short int *odata;
+ int i, j;
+ FILE *fp;
+ int size = tile_comp->width * tile_comp->height;
+ image = (float *) malloc(size * sizeof(float));
+ cuda_memcpy_dth(tile_comp->img_data_d, image, size * sizeof(float));
+
+ odata = (short int*) malloc(size * sizeof(short int));
+
+
+ for (j = 0; j < tile_comp->height; j++) {
+ for (i = 0; i < tile_comp->width; i++) {
+ odata[i + j * tile_comp->width] = (short int) image[i + j * tile_comp->width];
+ }
+ }
+
+
+
+ if ((fp = fopen(filename, "wb")) == NULL) {
+ printf("ERROR %d. Can not write file: %s \n", errno, filename);
+ system("PAUSE");
+ exit(1);
+ } else {
+ fseek(fp, 0L, SEEK_SET);
+ fwrite(odata, sizeof(short int), size * sizeof(short int), fp);
+ }
+ fclose(fp);
+ free(odata);
+ free(image);
+}
+
+void save_tile_comp_with_shift(type_tile_comp *tile_comp, char *filename, int shift)
+{
+ float *odata;
+ int i, j;
+ FILE *fp;
+ int size = tile_comp->width * tile_comp->height * sizeof(float);
+ odata = (float *) malloc(size);
+ cuda_memcpy_dth(tile_comp->img_data_d, odata, size);
+
+
+ for (j = 0; j < tile_comp->height; j++) {
+ for (i = 0; i < tile_comp->width; i++) {
+ odata[i + j * tile_comp->width] += 1 << (shift - 1);
+ printf("%.1f,", odata[i + j * tile_comp->width]);
+ }
+ printf("\n");
+ }
+
+
+ if ((fp = fopen(filename, "wb")) == NULL) {
+ printf("ERROR %d. Can not open file: %s \n", errno, filename);
+ system("PAUSE");
+ exit(1);
+ } else {
+ fseek(fp, 0L, SEEK_SET);
+ fwrite(odata, sizeof(float), size, fp);
+ }
+ fclose(fp);
+ free(odata);
+}
+
+int save_img_ord(type_image *img, const char *filename)
+{
+
+ int i, c, x, y;
+ type_tile *tile;
+ type_tile_comp *tile_comp;
+
+ int scan_width = img->width * (img->num_components);
+ BYTE *bits = (BYTE*) malloc(img->height * scan_width * sizeof(BYTE));
+ FIBITMAP *dst;
+ println_start(INFO);
+
+ // printf("for\n");
+ //exact opposite procedure as in read_img()
+ for (i = 0; i < img->num_tiles; i++) {
+ tile = &(img->tile[i]);
+ for (c = 0; c < img->num_components; c++) {
+ tile_comp = &(tile->tile_comp[c]);
+ cuda_h_allocate_mem((void **) &(tile_comp->img_data), tile_comp->width * tile_comp->height
+ * sizeof(type_data));
+ cuda_memcpy_dth(tile_comp->img_data_d, tile_comp->img_data, tile_comp->width * tile_comp->height
+ * sizeof(type_data));
+ for (x = 0; x < tile_comp->width; x++) {
+ for (y = 0; y < tile_comp->height; y++) {
+ bits[(tile->tly + y) * scan_width + (tile->tlx + x) * img->num_components + c]
+ = (BYTE) tile_comp->img_data[x + y * tile_comp->width];
+
+ // if(tile_comp->img_data[x + y * tile_comp->width] > (BYTE) tile_comp->img_data[x + y * tile_comp->width]) {
+ // printf("%f = %i\n", tile_comp->img_data[x + y * tile_comp->width], bits[(tile->tly + y) * scan_width + (tile->tlx + x) * img->num_components + c]);
+ // }
+ }
+ }
+ }
+ }
+
+ // convert the bitmap to raw bits (top-left pixel first)
+ dst = FreeImage_ConvertFromRawBits(bits, img->width, img->height, scan_width, 24, FI_RGBA_RED_MASK,
+ FI_RGBA_GREEN_MASK, FI_RGBA_BLUE_MASK, TRUE);
+
+ if (FreeImage_Save(FIF_BMP, dst, filename, 0)) {
+ println_var(INFO, "saved: %s", filename);
+ } else {
+ println_var(INFO, "saving FAILED: %s", filename);
+ }
+
+ FreeImage_Unload(dst);
+ free(bits);
+
+ return 0;
+
+}
+
+int save_img_grayscale(type_image *img, char *filename)
+{
+
+ int i, c, x, y;
+ type_tile *tile;
+ type_tile_comp *tile_comp;
+ int scan_width;
+ BYTE *bits;
+ FIBITMAP *dst;
+ println_start(INFO);
+
+ tile = &(img->tile[0]);
+ tile_comp = &(tile->tile_comp[0]);
+ cuda_h_allocate_mem((void **) &(tile_comp->img_data), tile_comp->width * tile_comp->height
+ * sizeof(type_data));
+ cuda_memcpy_dth(tile_comp->img_data_d, tile_comp->img_data, tile_comp->width * tile_comp->height
+ * sizeof(type_data));
+
+ scan_width = img->width * (img->num_components);
+ bits = (BYTE*) malloc(img->height * scan_width * sizeof(BYTE));
+
+ // printf("for\n");
+ //exact opposite procedure as in read_img()
+ for (i = 0; i < img->num_tiles; i++) {
+ for (c = 0; c < img->num_components; c++) {
+ for (x = 0; x < tile_comp->width; x++) {
+ for (y = 0; y < tile_comp->height; y++) {
+ bits[(tile->tly + y) * scan_width + (tile->tlx + x) * img->num_components + c]
+ = (BYTE) tile_comp->img_data[x + y * tile_comp->width];
+
+ // if(tile_comp->img_data[x + y * tile_comp->width] > (BYTE) tile_comp->img_data[x + y * tile_comp->width]) {
+ // printf("%f = %i\n", tile_comp->img_data[x + y * tile_comp->width], bits[(tile->tly + y) * scan_width + (tile->tlx + x) * img->num_components + c]);
+ // }
+ }
+ }
+ }
+ }
+
+ // convert the bitmap to raw bits (top-left pixel first)
+ dst = FreeImage_ConvertFromRawBits(bits, img->width, img->height, scan_width, 24, FI_RGBA_RED_MASK,
+ FI_RGBA_GREEN_MASK, FI_RGBA_BLUE_MASK, TRUE);
+
+ if (FreeImage_Save(FIF_BMP, dst, filename, 0)) {
+ println_var(INFO, "saved: %s", filename);
+ } else {
+ println_var(INFO, "saving FAILED: %s", filename);
+ }
+
+ FreeImage_Unload(dst);
+ free(bits);
+
+ return 0;
+
+}
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment